cuda - 对网格和 block 尺寸的混淆

Question

我正在尝试在 Udacity 的第 1 课结束时解决问题当然，但我不确定我是否只是打错字或者实际代码是否错误。

void your_rgba_to_greyscale(const uchar4 * const h_rgbaImage, uchar4 * const d_rgbaImage, unsigned char* const d_greyImage, size_t numRows, size_t numCols)
{
    size_t totalPixels = numRows * numCols;
    size_t gridRows = totalPixels / 32;
    size_t gridCols = totalPixels / 32;
    const dim3 blockSize(32,32,1);
    const dim3 gridSize(gridCols,gridRows,1);
    rgba_to_greyscale<<<gridSize, blockSize>>>(d_rgbaImage, d_greyImage, numRows, numCols);
    cudaDeviceSynchronize(); checkCudaErrors(cudaGetLastError());
}

另一种方法是:

void rgba_to_greyscale(const uchar4* const rgbaImage, unsigned char* const greyImage, int numRows, int numCols)
{   
    int x = (blockIdx.x * blockDim.x) + threadIdx.x;
    int y = (blockIdx.y * blockDim.y) + threadIdx.y;
    uchar4 rgba = rgbaImage[x * numCols + y];
    float channelSum = 0.299f * rgba.x + 0.587f * rgba.y + 0.114f * rgba.z;
    greyImage[x * numCols + y] = channelSum;
}

错误消息显示以下内容:

libdc1394 error: failed to initialize libdc1394
Cuda error at student_func.cu:76
unspecified launch failure cudaGetLastError()
we were unable to execute your code. Did you set the grid and/or block size correctly?

但是，它说代码已经编译，

Your code compiled!
error output: libdc1394 error: Failed to initialize libdc1394
Cuda error at student_func.cu:76
unspecified launch failure cudaGetLastError()

第 76 行是第一个代码块中的最后一行，据我所知，我没有更改任何内容。第76行如下，

rgba_to_greyscale<<<gridSize, blockSize>>>(d_rgbaImage, d_greyImage, numRows, numCols);

我实际上找不到 cudaGetLastError() 的声明.
我主要关心的是我对设置网格/ block 尺寸的理解+第一种方法方法是否适合一维像素位置数组和我的线程之间的映射。
编辑:
我想我误解了什么。是 numRows垂直方向的像素数？是 numCols水平方向的像素？
我的 block 由 8 x 8 个线程组成，每个线程代表 1 个像素？如果是这样，我假设这就是为什么我在计算 gridRows 时必须除以 4 的原因。因为图像不是正方形的？我假设我也可以制作一个 2:1 列的 block :行？

编辑 2:
我只是试图改变我的 block ，使它成为 2:1 的比例，所以我可以划分 numRows和 numCol相同的数字，但现在在底部和侧面显示空白区域。为什么底部和侧面都有空白区域。我没有改变网格或 block 的 y 尺寸。

Answer 1

每个 block 处理 32*32 像素，并且有 (totalPixels/32) * (totalPixels/32) 个 block ，所以你处理 totalPixels ^ 2 个像素 - 这似乎是错误的

第一个是错误的，这应该是正确的:

const dim3 blockSize(32,32,1);

size_t gridCols = (numCols + blockSize.x - 1) / blockSize.x;
size_t gridRows = (numRows + blockSize.y - 1) / blockSize.y;

这是 2d 的一个非常常见的模式 - 你可以记住它

在示例图像大小不是 2 的幂，并且您希望 block 覆盖所有图像(甚至更多)

所以下一个必须是正确的:
gridCols * blockSize.x >= numCols
gridRows * blockSize.y >= numRows

您选择 block 大小并根据它计算需要覆盖所有图像的 block 数量

之后，在内核中，您必须检查您是否没有“超出图像”，对于大小错误的情况

另一个问题在内核中，它必须是 (y * numCols + x)，而不是相反

核心:

int x = (blockIdx.x * blockDim.x) + threadIdx.x;
int y = (blockIdx.y * blockDim.y) + threadIdx.y;

if(x < numCols && y < numRows)
{
    uchar4 rgba = rgbaImage[y * numCols + x];
    float channelSum = 0.299f * rgba.x + 0.587f * rgba.y + 0.114f * rgba.z;
    greyImage[y * numCols + x] = channelSum;
}

调用代码:

const dim3 blockSize(4,32,1); // may be any

size_t gridCols = (numCols + blockSize.x - 1) / blockSize.x;
size_t gridRows = (numRows + blockSize.y - 1) / blockSize.y;

const dim3 gridSize(gridCols,gridRows,1);
rgba_to_greyscale<<<gridSize, blockSize>>>(d_rgbaImage, d_greyImage, numRows, numCols);
cudaDeviceSynchronize(); 
checkCudaErrors(cudaGetLastError());

该死，我觉得我做的事情更难理解(