CUDA 中的图像序列处理

Question

我的任务是使用 CUDA 实现图像重建算法代码。我获得了相同的 C 代码。代码的输入是一个 DAT 文件，其中包含 360 个大小为 640 x 480 的图像。代码如下所示：

    FILE *in,*out;
    float *i_data,*o_data;
    i_data=(float *)malloc(mem_size);
    for(int projection=0;projection<360;projection++)
    {
      in=fopen("filename.dat","rb");
      fread(i_data,mem_size,1,in);
      ... some math ...
      for(int slice_no=-240;slice_no<240:slice_no++)
      {
          for (i=-320;i<320;i++)
          for (j=-320;j<320;j++)
          {
             // do some operations
             (*(o_data*slice_no)+(j+320)+(i+240))+=(*(i_data*value)+(j+240)+(i+320));
             // some more math
          }
      } 
    }

输出浮点指针被写回一个 dat 文件。如果我必须并行化这些循环，我将如何在 CUDA 中做到这一点？我尝试使用 640 个线程中的每个 640 个块在 CUDA 中实现内部两个 for 循环。如何将线程索引赋予循环内的指针操作。我试着给

         int i=blockIdx.x;
         int j=threadIdx.x;

     and 

         kernel<<<640,640>>>

但这在输出指针中给出了错误的值。大多数是南。除了上面sn-p中显示的带有指针的行之外，我能够成功地实现其他数学。

谁能帮我做这件事？我在这段代码中做错了什么？ 也可以在这里并行化所有的for循环吗？

Answer 1

将内部循环限制在单个块内运行效率不高。此外，如果您并行化所有三个循环（如果您可以从磁盘加载更多数据并一次将许多图像发送到 gpu，则为 4 个循环），您将获得更多好处。

你的问题看起来像

for(x = minx; x < maxx; ++x)
{
  for(y = miny; y < maxy; ++y)
  {
    for(z = minz; z < maxz; ++z)
    {
      // do some math
    }
   }
}

内核中的线程数应该是：

num_thread = (maxx-minx) * (maxy-miny) * (maxz-minz);

你应该独立于块的大小。将其设置为blockDim.x = 256 之类的常量（尝试找到更好的常量）。

区块的数量将是

gridDim.x =(num_thread + blockDim.x) / blockDim.x

请注意，我的推理好像三个循环展开为一个大循环，但如果您的驱动程序允许，您可以制作 2D 甚至 3D 线程块。

在您的内核中，通过内置变量，您可以将大展开循环内的全局索引计算为

int index = blockIdx.x * blockDim.x + threadIdx.x

确保您没有超出范围

if(index >= num_thread) return; // do nothing

现在使用index，您可以恢复x、y和z（slice_no、i、j）为

x = index / ( (maxy-miny) * (maxz-minz) ) + minx;
y = ( index % ( (maxy-miny) * (maxz-minz) ) ) / (maxz-minz) + miny;
z = ( index % ( (maxy-miny) * (maxz-minz) ) ) % (maxz-minz) + minz;