3

我有以下 OpenCL 内核，一个高斯模糊

__constant sampler_t sampler =
        CLK_NORMALIZED_COORDS_FALSE |
        CLK_ADDRESS_CLAMP_TO_EDGE |
        CLK_FILTER_NEAREST;

__constant float gaussian_kernel[3][3] = {
    {0.0625f, 0.125f, 0.0625f},
    {0.125f, 0.25f, 0.125f},
    {0.0625f, 0.125f, 0.0625f} };

void kernel gaussian_blur(
    read_only image2d_t input_image,
    write_only image2d_t output_image) {

    int x = get_global_id(0);
    int y = get_global_id(1);

    int2 coords[9] = {
        { x - 1, y - 1 }, { x, y - 1 }, { x + 1, y - 1 },
        { x - 1, y     }, { x, y     }, { x + 1, y     },
        { x - 1, y + 1 }, { x, y + 1 }, { x + 1, y + 1 }
    };

    float4 pixel_value = { 0.f, 0.f, 0.f, 0.f };

    for(int i = 0; i < 3; ++i) {
        for(int j = 0; j < 3; ++j) {
            int index = i * 3 + j;
            float4 blurred =
                as_float4(read_imageui(input_image, sampler, coords[index]));

            pixel_value.x += (blurred.x * gaussian_kernel[i][j]);
            pixel_value.y += (blurred.y * gaussian_kernel[i][j]);
            pixel_value.z += (blurred.z * gaussian_kernel[i][j]);
            pixel_value.w += (blurred.w * gaussian_kernel[i][j]);
        }
    }

    uint4 final_value = as_uint4(pixel_value);

    write_imageui(output_image, coords[4], final_value);
}

当我指定设备用作 CPU 时，模糊工作正常。这是设备选择代码

std::vector<cl::Platform> all_platforms;
cl::Platform::get(&all_platforms);

if(all_platforms.size() == 0) {
    std::cerr << "No platforms available" <<std::endl;
    exit(-1);
}

cl::Platform default_platform = all_platforms[0];

std::vector<cl::Device> all_devices;
default_platform.getDevices(CL_DEVICE_TYPE_ALL, &all_devices);

if(all_devices.size() == 0) {
    std::cerr << "No device found" << std::endl;
    exit(-1);
}

cl::Device default_device = all_devices[1]; //Changing this index to 0 uses my graphics card

现在，如果将 default_device 设置为 GPU，则程序只会输出一个空图像。相关的图像设置代码是（注意这input是一个Magick::Image和in_pixels一个堆分配的数组unsigned short）：

cl::ImageFormat format(CL_RGBA, CL_UNSIGNED_INT16);

cl::Image2D input_image_buffer;
input.write(0, 0, 
    input.baseColumns(), input.baseRows(), "RGBA", Magick::ShortPixel, in_pixels);

input_image_buffer = cl::Image2D(
    context, 
    CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
    format,
    input.baseColumns(),
    input.baseRows(),
    0,
    in_pixels,
    &cl_error);

cl::Image2D output_image_buffer;
output_image_buffer = cl::Image2D(
    context, 
    CL_MEM_WRITE_ONLY | CL_MEM_USE_HOST_PTR,
    format, 
    input.baseColumns(),
    input.baseRows(),
    0,
    out_pixels,
    &cl_error);

并且内核设置/图像输出代码（gaussian_program当然是没有错误地构建的）

cl::Kernel gaussian_kernel(gaussian_program, "gaussian_blur");

cl::CommandQueue queue(context, default_device, 0, &cl_error);

cl::size_t<3> origin;
cl::size_t<3> size;
origin[0] = 0;
origin[1] = 0;
origin[2] = 0;

size[0] = input.baseColumns();
size[1] = input.baseRows();
size[2] = 1;

cl_error = gaussian_kernel.setArg(0, input_image_buffer);

cl_error = gaussian_kernel.setArg(1, output_image_buffer);

cl::NDRange range(input.baseColumns(), input.baseRows());

cl_error = queue.enqueueNDRangeKernel(
    gaussian_kernel, 
    cl::NullRange, 
    range,
    cl::NullRange);

queue.finish();

try{
    output.read(
        input.baseColumns(), 
        input.baseRows(), 
        "RGBA", Magick::ShortPixel, out_pixels);
}
catch(Magick::Exception& ex) {
    std::cerr << "A Magick error occured while writing the pixel cache: " <<
        std::endl << ex.what() << std::endl;
    return false;
}

现在，出于本示例的目的，我删除了许多错误检查，但原始代码在每次 OpenCL 调用后检查 cl_error 并且从不发出错误信号。代码在 CPU 上按预期执行，但在 GPU 上执行代码时图像为空。

起初我怀疑存在同步问题（该queue.finish()调用是为了实现该精确目的而需要的，即使在 CPU 上也是如此），但是为了序列化执行而在代码中乱扔cl::finish()或queue.finish()调用根本没有帮助。

有什么我明显做错了吗？这个 OpenCL 内核是否有可能在 GPU 上失败但在 CPU 上失败？

作为记录，我在 Ubuntu 13.04 上使用 AMD APP SDK OpenCL 实现和 Radeon HD 7970。

4

2 回答 2

6

那么，你在哪里从你的 GPU 读回图像？

您使用标志分配输入图像CL_MEM_COPY_HOST_PTR

仅当 host_ptr 不为 NULL 时，此标志才有效。如果指定，则表明应用程序希望 OpenCL 实现为内存对象分配内存并从 host_ptr 引用的内存中复制数据。

和你的输出图像CL_MEM_USE_HOST_PTR

仅当 host_ptr 不为 NULL 时，此标志才有效。如果指定，则表明应用程序希望 OpenCL 实现使用 host_ptr 引用的内存作为内存对象的存储位。

OpenCL 实现允许将 host_ptr 指向的缓冲区内容缓存在设备内存中。当内核在设备上执行时，可以使用此缓存副本。

对使用相同 host_ptr 或重叠主机区域创建的多个缓冲区对象进行操作的 OpenCL 命令的结果被认为是未定义的。

您的分配没有任何问题，但是您永远不会告诉 OpenCL 实现将内存写入使用的设备并将该内存读回主机主内存。这可能适用于 CPU，因为内存已经在它们的设备内存（主内存）中，但对于 GPU 来说却失败了。

OpenCL c++ 绑定提供cl::enqueueWriteImage(/*params*/);和cl::enqueueReadImage(/*params*/);向设备写入和读取图像缓冲区。

于 2013-09-13T06:06:40.293 回答
2

正如我的评论中所述，as_float4不是转换。它需要uint的 32 位，并将它们解释为浮点位。在您的情况下，您正在读取 16 位，因此浮点值将非常小（指数将为 0）。请改用convert_float4。

关于不读回数据的答案也是正确的。您需要调用clEnqueueReadBuffer或clEnqueueMapBuffer以确保从设备读回数据。

于 2013-09-13T21:18:17.373 回答