c++ - 简单的 OpenCL 内核在输出图像中生成垂直条而不是纯色。为什么？

Question

我正在尝试编写图像处理 OpenCL 应用程序，但我的问题是任何更改输入图像的尝试都会产生看起来像垂直条的伪像。如果我复制图像像素而不改变它们，就不会发生这种情况。因此，例如，这条线会产生工件:

pixel = (uint4)(image1_pixel.x,
                image1_pixel.y,
                image1_pixel.z,
                255);

...但这一个按预期工作:

pixel = (uint4)(image1_pixel.x,
                image1_pixel.y,
                image1_pixel.z,
                image1_pixel.w);

Input是不透明的 32 位 PNG 图像，因此我希望两行代码产生相同的结果。然而，实际上，只有第二行按预期工作。第一行给出 output with artifacts .

这是我的内核:

__constant sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE |
                              CLK_ADDRESS_CLAMP |
                              CLK_FILTER_NEAREST;

__kernel void test(__read_only image2d_t image1,
                  __write_only image2d_t out) {
  const int2 pos = (int2)(get_global_id(0), get_global_id(1) );
  uint4 image1_pixel = read_imageui(image1, sampler, pos);
  uint4 pixel = (uint4)(image1_pixel.x,
                        image1_pixel.y,
                        image1_pixel.z,
                        255);
  write_imageui(out, pos, pixel);
}

这是 main.cpp 代码的相关部分:

  CImg<unsigned char> image1("../input.png");
  ...
  Image2D clImage1 = Image2D(context,
    CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
    ImageFormat(CL_RGBA, CL_UNSIGNED_INT8),
    image1.width(), image1.height(), 0, image1.data() );
  Image2D clResult = Image2D(context, CL_MEM_WRITE_ONLY,
    ImageFormat(CL_RGBA, CL_UNSIGNED_INT8),
    image1.width(), image1.height(), 0, NULL);
  Kernel test = Kernel(program, "test");
  test.setArg(0, clImage1); test.setArg(1, clResult);
  Event kernel_event, read_event;
  queue.enqueueNDRangeKernel(test, NullRange,
    NDRange(image1.width(), image1.height() ),
    NullRange, NULL, &kernel_event);
  cl::size_t<3> origin;
  origin.push_back(0); origin.push_back(0); origin.push_back(0);
  cl::size_t<3> region;
  region.push_back(image1.width() );
  region.push_back(image1.height() ); region.push_back(1);
  queue.enqueueReadImage(clResult, CL_TRUE,
                         origin, region, 0, 0,
                         image1.data(), NULL, NULL);
  kernel_event.wait();
  image1.save("../output.png");

Here可以下载我的测试应用程序的完整源代码(它包含 30 行以下的简短 main.cpp、CMakeLists.txt、解释如何编译和运行它的 readme.txt、输入图像和内核)。我使用 CImg 库加载和保存图像。我仔细检查了输入是否以 32 位 RGBA 图像打开。我尝试用 AMD 或 NVidia SDK 运行内核并得到相同的结果。

知道为什么我会得到意想不到的结果吗？

Answer 1

CImg使用平面格式，而OpenCL的image2d_t期望interleaved (例如平面是:R1R2R3R4R5R6...G1G2G3G4G5G6...B1B2B3B4B5B6...，交错是:R1G1B1R2G2B2R3G3B3...) 当您复制像素而不更改它们时，平面格式不会丢失并且可以正常工作，但是如果您开始修改图像的一个组件，您实际上会修改像素的所有组件，从而修改黑度，因为这个平面/交错问题。

如果你想坚持使用 CImg，你可以在将图像发送到设备之前对其进行多路复用:

CImg<unsigned int> muxRGBA(const CImg<unsigned int>& pic)
{
    assert( pic.spectrum() == 3 );
    CImg<float> mux(pic.width()*4,pic.height(),pic.depth(),1);

    cimg_forXYZ(pic,x,y,z){    
        for(int k=0; k<3; k++) {
            mux(x*4+k,y,z,0) = pic(x,y,z,k);
        }
        mux(x*4+3,y,z,0) = 1.0f; // alpha channel
    }
    return mux;
}

然后在处理后解复用，所以你仍然可以用CImg显示/保存它:

CImg<unsigned int> demuxRGBA(const CImg<unsigned int>& pic)
{
    assert( pic.spectrum()==1 );
    assert( pic.width()%4 == 0 );
    CImg<float> demux(pic.width()/4,pic.height(),pic.depth(),3);

    cimg_forXYZ(demux,x,y,z) {
        for(int k=0; k<3; k++) {
            demux(x,y,z,k) = pic(x*4+k,y,z,0);
        }
    }
    return demux;
}