如何在 CUDA 中正确添加全局内存？

Question

我正在尝试在 CUDA 中为家庭作业实施绝对差异之和，但无法获得正确的结果。 我得到一个 Blocksize，它代表我要比较的图像的正方形部分的 X 和 Y 大小（以像素为单位）。我还得到了两张 YUV 格式的图像。以下是我必须实现的程序部分：计算 SAD 的内核和线程网格/块大小的设置。程序的其余部分已提供，可以假定是正确的。

在这里，我获取当前线程的 x 和 y 索引，并使用它们来获取我在当前线程中处理的图像数组中的像素。然后我计算绝对差，等待所有线程完成计算，然后如果当前线程在图像中的块内，我们关心的绝对差将通过 atomicAdd 添加到全局内存中的总和以避免碰撞在写入期间。

__global__ void gpuCounterKernel(pixel* cuda_curBlock, pixel* cuda_refBlock, uint32* cuda_SAD,  uint32 cuda_Blocksize)                                                                                                                  
 {                                                                                                                                                                                                                                                                                                                                                               
     int idx = blockIdx.x * blockDim.x + threadIdx.x;                                                                                                                                                                                    
     int idy = blockIdx.y * blockDim.y + threadIdx.y;                                                                                                                                                                                    
     int id = idx * cuda_Blocksize + idy;                                                                                                                                                                               
     int AD = abs( cuda_curBlock[id] - cuda_refBlock[id] );                                                                                                     
     __syncthreads();                                                                                                                                                                                 
     if( idx < cuda_Blocksize && idy < cuda_Blocksize ) {                                                                                                                                                                           
         atomicAdd( cuda_SAD, AD );                                                                                                                                                                                              
     }                                                                                                                                                                                                                                   
 }

这就是我为内核设置网格和块的方式：

int grid_sizeX   = Blocksize/2;                                                                                                                                                                                                
int grid_sizeY   = Blocksize/2;                                                                                                                                                                                                     
int block_sizeX  = Blocksize/4;                                                                                                                                                                                                     
int block_sizeY  = Blocksize/4;
dim3 blocksInGrid(grid_sizeX, grid_sizeY);                                                                                                                                                                                     
dim3 threadsInBlock(block_sizeX, block_sizeY);                                                                                                                                                                              

给定的程序也会计算 CPU 上的 SAD，并将我们从 GPU 得到的结果与那个结果进行比较以检查正确性。图像中的有效块大小为 1-1000。我上面的解决方案是从 10-91 得到正确的结果，但任何高于 91 的结果都只返回 0 作为总和。我做错了什么？

Answer 1

您的网格和块大小设置看起来很奇怪。

通常我们使用类似如下的图像像素设置。

int imageROISize=1000;
dim3 threadInBlock(16,16);
dim3 blocksInGrid((imageROISize+15)/16, (imageROISize+15)/16);

您可以参考 cuda 编程指南中的以下部分，了解有关如何将工作负载分配到 CUDA 线程的更多信息。

http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#thread-hierarchy

Answer 2

您确实应该显示所有代码并确定您正在运行的 GPU。至少是调用内核并分配数据供 GPU 使用的部分。

• 您是否在所有 cuda API 调用和内核调用上都使用了正确的cuda error checking？

• 可能你的内核根本没有运行，因为你的 threadsInBlock 参数总共超过 512 个线程。您指出在Blocksize = 92 及以上，事情不工作。让我们算一下：

  92/4 = 23 threads in X and Y dimensions
  23 * 23 = 529 total threads requested per threadblock
  

529 超过 512，这是 cc 1.x 设备的限制，所以我猜你是在 cc 1.x 设备上运行，因此你的内核启动失败，所以你的内核没有运行，并且所以你没有得到计算结果（即0）。请注意，在 X 和 Y 维度上 91/4 = 22 个线程时，您请求的线程总数为 484 个 does not exceed the 512 limit for cc 1.x devices。

如果您进行了正确的 cuda 错误检查，错误报告会将您的注意力集中在由于启动参数不正确而导致 cuda 内核启动失败。