如何制作 CUDA 直方图内核？答案

【问题标题】：how to make a CUDA Histogram kernel?如何制作 CUDA 直方图内核？
【发布时间】：2011-07-02 05:11:00
【问题描述】：

我正在为图片上的直方图编写一个 CUDA 内核，但我不知道如何从内核返回一个数组，当其他线程读取它时，该数组会改变。有什么可能的解决方案吗？

__global__ void Hist(
    TColor *dst, //input image
    int imageW,
    int imageH,
 int*data
){
    const int ix = blockDim.x * blockIdx.x + threadIdx.x;
    const int iy = blockDim.y * blockIdx.y + threadIdx.y;

if(ix < imageW && iy < imageH)
{
  int pixel = get_red(dst[imageW * (iy) + (ix)]);
                  //this assign specific RED value of image to pixel

  data[pixel] ++; // ?? problem statement ...
 }
}

@para d_dst：输入图像TColor等于float4。

@para data：直方图大小的数组[255]

extern "C" void
cuda_Hist(TColor *d_dst, int imageW, int imageH,int* data) 
{
  dim3 threads(BLOCKDIM_X, BLOCKDIM_Y);
  dim3 grid(iDivUp(imageW, BLOCKDIM_X), iDivUp(imageH, BLOCKDIM_Y));
  Hist<<<grid, threads>>>(d_dst, imageW, imageH, data);
}

【问题讨论】：

标签： histogram cuda

【解决方案1】：

您查看过 SDK 示例吗？ CUDA SDK 中提供了“直方图”示例（NVIDIA developer site 上的当前版本为 3.0，registered developers 上的版本为 3.1 beta）。

带有示例的文档很好地解释了如何处理求和，或者使用 GPU 上的全局内存原子，或者分别收集每个块的结果，然后进行单独的归约（在主机或 GPU 上）。

【讨论】：

【解决方案2】：

在使用 CUDA（或通常使用 GPGPU）实现时，直方图并不是特别有效 - 通常您需要在共享内存中生成大量部分直方图，然后对它们求和。您可能需要考虑将此特定任务保留在 CPU 上。

【讨论】：

但是，我的任务是尝试使用 CUDA 来应用直方图。我无法完成它。数据无法单独实现

【解决方案3】：

您将不得不使用原子函数来阻止其他线程使用相同的内存，或者使用部分直方图。无论哪种方式，除非输入图像非常大，否则效率不高。

【讨论】：