3D 图像索引

Question

我有一张尺寸为 512 x 512 x 512 的图像。 我需要单独处理所有体素。 我怎样才能获得线程ID来做到这一点？ 如果我使用 1D 线程 ID，块数将超过 65536。

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

注意：- 我的卡不支持 3D 网格

Answer 1

您可以在 CUDA 4.0 和计算能力 2.0+ 中使用 3D 索引。示例代码：

int blocksInX = (nx+8-1)/8;
int blocksInY = (ny+8-1)/8;
int blocksInZ = (nz+8-1)/8;

dim3 Dg(blocksInX, blocksInY, blocksInZ);
dim3 Db(8, 8, 8);
foo_kernel<<Dg, Db>>(R, nx, ny, nz);

...

__global__ void foo_kernel( float* R, const int nx, const int ny, const int nz )
{
  unsigned int xIndex = blockDim.x * blockIdx.x + threadIdx.x;
  unsigned int yIndex = blockDim.y * blockIdx.y + threadIdx.y;
  unsigned int zIndex = blockDim.z * blockIdx.z + threadIdx.z;

  if ( (xIndex < nx) && (yIndex < ny) && (zIndex < nz) )
  {
    unsigned int index_out = xIndex + nx*yIndex + nx*ny*zIndex;
    ...
    R[index_out] = ...;
  }
}

如果您的设备不支持计算能力 2.0，则有一些技巧：

int threadsInX = 16;
int threadsInY = 4;
int threadsInZ = 4;

int blocksInX = (nx+threadsInX-1)/threadsInX;
int blocksInY = (ny+threadsInY-1)/threadsInY;
int blocksInZ = (nz+threadsInZ-1)/threadsInZ;

dim3 Dg = dim3(blocksInX, blocksInY*blocksInZ);
dim3 Db = dim3(threadsInX, threadsInY, threadsInZ);

foo_kernel<<<Dg, Db>>>(R, nx, ny, nz, blocksInY, 1.0f/(float)blocksInY);

__global__ void foo_kernel(float *R, const int nx, const int ny, const int nz,
                           unsigned int blocksInY, float invBlocksInY)
{

    unsigned int blockIdxz = __float2uint_rd(blockIdx.y * invBlocksInY);
    unsigned int blockIdxy = blockIdx.y - __umul24(blockIdxz, blocksInY);
    unsigned int xIndex = __umul24(blockIdx.x, blockDim.x) + threadIdx.x;
    unsigned int yIndex = __umul24(blockIdxy, blockDim.y) + threadIdx.y;
    unsigned int zIndex = __umul24(blockIdxz, blockDim.z) + threadIdx.z;

    if ( (xIndex < nx) && (yIndex < xIndex) && (zIndex < nz) )
    {
        unsigned int index = xIndex + nx*yIndex + nx*ny*zIndex;
        ...
        R[index] = ...;
    }

}

Answer 2

你可以使用网格。它为您提供了更多索引。

Answer 3

请注意，您的 PC 内存不是 3D 的。这只是可视化的问题，因此您可以将 3D 图像转换为单个指针。

Array[i][j][z] is same as Array2[ i*cols+j + rows*cols*z];

现在将 Array2 馈送到 CUDA 并在单维中工作

Answer 4

如果您需要更大的网格，CUDA 在所有硬件上支持 2D 网格，并且最新版本的 CUDA 工具包还支持当前 Fermi 硬件上的 3D 网格。

然而，并不是绝对有必要拥有这么大的网格。如果每个体素操作都是独立的，那么为什么不只使用一维网格，而是让每个线程处理多个体素呢？这样的方案不仅不需要更大的 2D 或 3D 网格，而且可能会更有效，因为与块的调度和初始化相关的固定成本可以通过多个体素计算进行摊销。

Answer 5

我使用了这样的东西：

在代码中定义你的网格：
暗淡 3 替代网格，替代线程；
altgrid.x=lx;
altgrid.y=ly;
altgrid.z=1；
altthreads.x=lz;
altthreads.y=1;
altthreads.z​​=1；

并且在内核中

int idx = threadIdx.x;
int idy = blockIdx.x ;
int idz = blockIdx.y ;

由于设备上的数组仅为 1D，因此您可以通过矩阵 A 检索 [idx][idy][idz] 元素作为 A[ind]，其中 ind=idz+lz*(idy+ly*idx ）；

我希望它有帮助