处理图像时出现 CUDA 错误

Question

我正在尝试将黑白图像作为输出，并将彩色图像作为输入。我正在使用 OpenCV 来获取图像并写入输出，并使用 CUDA 在内核中制作黑白图像。我尝试了相同的代码，但没有使用 OpenCV，它运行良好。但现在的输出与我真正期望得到的略有不同。

我认为 CUDA 代码需要一些修改才能与 OpenCV 一起使用。我用它做了一些工作，但没有找到方法。也许有人可以给我一个建议或修改我的代码，好吗？我真的对这个问题感到困惑。

   __global__ void addMatrix(uchar4 *DataIn, unsigned char *DataOut)
    {
        int idx = blockIdx.x * blockDim.x + threadIdx.x;
        DataOut[idx] = (DataIn[idx].x + DataIn[idx].y + DataIn[idx].z)/3;
    }

int main() 
{
        cudaDeviceProp deviceProp;
        cudaGetDeviceProperties(&deviceProp, 0);

        char* c = "";
        printf("Input source of image\n Example of right directory file: E:\henrik-evensen-castle-valley-v03.jpg\n Your turn:\n");
        char *tbLEN;
        tbLEN = new char [1024];

        cin.getline(tbLEN,1024);

        cout<< endl << "Your image: " << tbLEN << endl;

        //Data for input image
        IplImage* image;
        image = cvLoadImage(tbLEN, 1);
        int height = image->height;
        int width = image->width;
        int step = image->widthStep;
        int SizeIn = (step*height);
        printf("\nProcessing image\n");
        //Data for output image
        IplImage *image2 = cvCreateImage(cvSize(width, height), IPL_DEPTH_8U, 1);
        int step2 = image2->widthStep;
        int SizeOut = step2 * height;

        //GPU
        uchar4* DatIn = (uchar4*)image->imageData;
        unsigned char* DatOut = (unsigned char*)image2->imageData;
        uchar4 *datIndev;
        unsigned char *datOutdev;

        printf("Allocating memory on Device\n");
        /* Allocate memory on Device */
        cudaMalloc(&datIndev, SizeIn * sizeof(unsigned char));
        cudaMalloc(&datOutdev, SizeOut * sizeof(unsigned char));

        printf("Copy data on Device\n");
        /* Copy data on Device */
        cudaMemcpy(datIndev, DatIn, SizeIn * sizeof(unsigned char), cudaMemcpyHostToDevice);
        cudaMemcpy(datOutdev, DatOut, SizeOut * sizeof(unsigned char), cudaMemcpyHostToDevice);

        int NumThreadsX = deviceProp.maxThreadsPerBlock;
        int NumBlocksX = (width * height)/NumThreadsX;

        dim3 blocks(NumBlocksX, 1, 1);
        dim3 threads(NumThreadsX, 1, 1);
        addMatrix <<< blocks, threads >>> (datIndev, datOutdev);


        cudaMemcpy(DatOut, datOutdev, SizeOut * sizeof(unsigned char), cudaMemcpyDeviceToHost);
        cvNamedWindow("Imagecolor");
        cvShowImage("Imagecolor", image);

        cvNamedWindow("Gray");
        cvShowImage("Gray", image2);
        const char* filename1 = "CcPwSwMW4AELPUc.jpg";
        printf("Saving an output image\n");
        cvSaveImage( filename1, image2 );
        cudaFree(datOutdev);
        cudaFree(datIndev);
        cvWaitKey(0);
        return 0;
}

Answer 1

这里有几个问题：

1. 您关于四通道数据的假设不正确。您的代码会将三通道 BGR 图像从文件加载到内存中。因此，您需要将引用从 uchar4 更改为 uchar，然后让每个线程从内核中的源图像加载三个字节

2. 您的内核本身包含一个潜在的算术错误。三个 unsigned char 像素值的总和可能会溢出 unsigned char 中间结果并产生不正确的平均值。您应该使用更大的类型进行计算。

综合起来，你的内核应该是这样的：

__global__ void addMatrix(unsigned char *DataIn, unsigned char *DataOut)
{
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    int b = DataIn[3*idx];
    int g = DataIn[3*idx+1];
    int r = DataIn[3*idx+2];

    DataOut[idx] = (unsigned char)((b + r + g)/3);
}

然后您可能会发现您的图像看起来正确。