0

我正在尝试在此代码中导入 CUDA：

double square=0;
for(int j=0;j<width; j++) {
  double Up=0,Down=0;
  for(int i=0;i<height; i++) {
    if(array1[i]>0 && array2[i]>0){
      square = source[i*width+j];
      square = square*square;
      Up   += square*array2[i]/array1[i];
      Down += square;
    }
  }
  if(Down>0){
    out[j] *= (1.+(Up/Down-1.));
  }
}

在第一次尝试中，我减少了第一个 for 循环。（效果很好）

int j = blockDim.x * blockIdx.x + threadIdx.x;

double Up=0, Down=0, square=0;
if (j<width) {
  for(int i=0;i<height;i++) {
    if(array1[i]>0 && array2[i]>0){
      square = source[i*width+j];
      square = square*square;
      Up   += square*array2[i]/array1[i];
      Down += square;
    }
  }
  if(Down>0){
    out[j] *= (1.+(Up/Down-1.));
  }
}

我也会减少第二个 for 循环，我用 2D 网格尝试过它不起作用。这是内核：

int j = blockDim.x * blockIdx.x + threadIdx.x;
int i = blockDim.y * blockIdx.y + threadIdx.y;
int offset = j + i * blockDim.x * gridDim.x;

double Up[width],Down[width], square[height];
if (j>=width && i>=height) return;

if(array1[i]>0 && array2[i]>0){
  square[i] = source[offset]*source[offset];
  Up[j]   += square[i]*array2[i]/array1[i];
  Down[j] += square[i];
}
if(Down[j]>0){
  out[j] *= (1.+(Up[j]/Down[j]-1.));
}

这是内核调用：

dim3 blocks(32,32);
dim3 grid(width/32,height/32);
kernel <<< grid, blocks >>> (...);
cudaDeviceSynchronize();

...错误是什么？有更有效的解决方案吗？（我可以使用动态并行？）

非常感谢！

4

1 回答 1

1

在您的最后一个内核中，您似乎打算将Up,Down和的数组square保留在线程之间，但是这些数组是线程本地的，因此它们包含的数据不在线程之间共享。不幸的是，即使它们在线程之间共享，您的方法也不起作用。

在您的内部循环中，当前循环使用上一轮计算的数据。并行化这样的循环并不是一件容易的事，有时它根本无法完成。在您的情况下，一个简单的解决方案是使用原子运算符来增加Up和Down计数器，但这不会有效，因为原子运算符会导致操作的隐式序列化。

您可能应该考虑使用已经优化的现有并行原语（例如前缀和）来解决这个问题。例如，CUB或Thrust中的那些。

于 2013-10-13T02:50:33.480 回答