1

如果多个线程同时写入一个内存位置。，会有竞争条件，对吧？？在我的情况下，同样的事情正在发生..

考虑来自“reduce.cl”的模块

int i = get_global_id(0);
int n,j;

n = keyMobj[i];                       // this n is the key..It can be either 0 or 1.
for(j=0; j<2; j++)
      sumMobj[n*2+j] += dataMobj[i].dattr[j];        //summing operation.

在这里，内存位置
sumMobj===> [...0..., ....1...] 被 4 个线程同时访问 & sumMobj===> [....3..., . ...4 ...]同时访问6个线程..

有没有办法让它并行，比如使用锁定或信号量？因为这个求和是我算法中非常重要的一部分......

4

1 回答 1

3

我可以给你一些提示，因为我也面临类似的问题。

我可以想到三种不同的方法来实现类似的目标：

考虑一个简单的内核，假设您启动了 4 (0-3) 个线程

_kernel void addition (int *p)
{
int i = get_local_id(0);
     p[4]+= p[i];
}

您希望将值 p[0]、p[1]、p[2]、p[3]、p[4] 相加，并将最终总和存储在 p[4] 中。正确的？IE：

p[4]= p[0] + p[1] + p[2] + p[3] + p[4]

方法-1 （无并行）

将此作业分配给仅 1 个线程（无并行性）：

int i = get_local_id(0);
if (i==0)

{

p[4]+= p[i];

}

方法 2（具有并行性）

表达你的问题如下：

p[4]= p[0] + p[1] + p[2] + p[3] + p[4] + 0

这是一个减少问题

所以启动 3 个线程：i=0 到 i=2。在第一次迭代中

 i=0 finds p[0] + p[1]
 i=1 finds p[2] + p[3]  
 i=2 finds p[4] + 0

现在你有了三个数字，你应用与上面相同的逻辑并添加这些数字（适当的填充 0 使其成为 2 的幂）

方法-3原子操作

如果您仍然需要以原子方式实现此功能，可以使用atomic_add()：

  int fsfunc atomic_add (   volatile __global int *p ,int val)

描述

读取存储在 p 指向的位置的 32 位值（称为旧值）。计算 (old + val) 并将结果存储在 p 指向的位置。该函数返回旧的。

这是假设数据是 int 类型。否则，您可以看到上面建议的链接。

于 2013-02-06T07:16:26.157 回答