opencl中原子锁的使用

opencl的核函数执行的是并行操作，因为每一个核函数执行的是一个细粒度的操作，所以当核函数执行的时候，相当于多个线程同时执行一个函数，区别就是这些线程输入给函数的参数不一样，比如对于一个图像10*10的图像，同时有100个线程操作，每一个线程操作一个像素，当然必须是不同的像素，所以不能有并发写的情况发生。

比如下面的核函数：

__kernel void sync_write(__global unsigned char * rgbImage, 
                              __global float * result)
{
    int x = get_global_id(0);
    int y = get_global_id(1);
    result[x-y] += 100;
}

我们对result数组进行写操作，因为对于不同的线程，x-y的值可能相同，比如x=0,y=0和x=100,y=100写入的是同一个地方，所以不能直接这样写。

opencl提供了原子锁的机制，只允许同时有一个线程进入：

atomic_add(result + y - x, 100);

第一个 参数是地址，第二个参数是要往这个地址增加的值。

atomic_add官方的解释如下：

这就是说，它只支持整型元素的相加，对于浮点型和其它的类型不支持

如果遇到浮点型的相加，我们可以通过乘法将浮点型转换为整型，然后在cpu里面再将整型数组转回去，这样的话还是比单纯的cpu运算要快