CUDA 内核的奇怪行为

Posted 2023-03-23

【中文标题】CUDA 内核的奇怪行为

【英文标题】：Strange behavior of CUDA kernel

【发布时间】：2021-12-10 04:58:14

【问题描述】：

我正在尝试制作一个简单的 Cuda 应用程序来创建给定矩阵的积分图像。我需要做的其中一个步骤是创建每一行的完整图像。为了做到这一点，我想为每一行分配 1 个线程。应该执行此操作的函数：

__global__ void IntegrateRows(const uchar* img, uchar* res)

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

    if (x >= Width || y >= Height)
        return;

    int sum = 0;
    int row = y * Width;
    for (int i = 0; i < Width - x; ++i)
    
        res[row + i + x] = sum + img[row + i + x];
        sum += img[row + i + x];
    

我使用一个大小为 3840x2160 的矩阵填充了一个 (cv::Mat::ones(Size(Width, Height), CV_8UC1)) 进行测试。当我尝试打印结果的内容时，它总是返回从 1 到 255 的数字序列：

执行配置为：

dim3 threadsPerBlock(1, 256);
dim3 numBlocks(1, 16);
IntegrateRows<<<numBlocks, threadsPerBlock >>>(img, res);

我的 GPU 是 Nvidia RTX 3090。

【问题讨论】：

res 是 unsigned char 类型。 255 是最大可以存储在 8 位类型。将 res 矩阵类型更改为更宽的 16 或 32 位整数。

OpenCV 矩阵使用倾斜内存 AFAIK。我非常怀疑您的索引是否正确，即使您修复了 8 位整数溢出错误

【参考方案1】：

tl;dr：让你的输出矩阵有更大的元素

如果你对序列进行积分/前缀求和

1, 1, 1, 1, ...

你得到：

0, 1, 2, 3, ...

当你达到元素类型的最大值时，这个序列将环绕到 0。在您的情况下，它是uchar，即unsigned char。它的最大值是 255。再加上一个 1，你得到 0。所以：0,1,2,3,... 253,254,255,0,1,...等等。

如果您将输出矩阵元素类型更改为unsigned short（或者可能只是unsigned int） - 您将不会获得环绕行为。当然，如果您将 255 而不是 1 相加，和/或您的矩阵更大，那么再次表示类型的范围可能不够大。