c++ - OpenCL 内核访问 CL_MEM_READ_WRITE，缓冲区

Question

我正在尝试将 float3 缓冲区 border 用于数据结构。它在使用英特尔 OpenCL SDK 4.4、英特尔 iCore7 的内核执行期间崩溃。不幸的是，我无法找出 3d-index (i, y, x) 到线性索引 adr=WIDTH2*i+WIDTH*y+x 中的任何索引错误然而。我错过了什么？

以下是缓冲区定义(使用 OpenCL C++ 包装器 v1.2):

m_numPixels(width*width),
m_inBuffer(getContext(), CL_MEM_READ_ONLY | CL_MEM_ALLOC_HOST_PTR, 
    sizeof(float)*(width*width), NULL),
m_inBuffer2(getContext(), CL_MEM_READ_ONLY | CL_MEM_ALLOC_HOST_PTR, 
    sizeof(float)*(width*width), NULL),
m_backBuffer(getContext(), CL_MEM_READ_ONLY, 
     sizeof(float)*(width*width), NULL),
m_borderBuffer(getContext(), CL_MEM_READ_WRITE, 
   (3*sizeof(float))*(10*width*width), NULL),
m_outBuffer(getContext(), CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR, 
    4*(width*width), NULL),

内核代码如下:

__kernel void computeMedial (__global const float* height, // input height
                 __global const float* height2, // input height, previous frame
                 __global const float* background, // input background
                 __global float3*      border,  // border datastructure
                 __global uchar4*      output,  // output image
                 float thres,
                 uint  width,
                 uint  ls,
                 float scale)
{
    uint   x = get_global_id(0);
    uint   y = get_global_id(1); 
    const uint WIDTH2 = width*width;
    const uint WIDTH  = width;

    // access pixel (x, y)
    float2 c00 = (float2)(x, y);
    float  h00 = array_height(height, height2, width, c00);
    if (x < 4 || x > width-5 || y < 4 || y > width-5) { // border location
      return;
    }
    if (h00 < thres-10 || h00 > thres+10) { // not in thres interval
      return;
    }
    output[y*WIDTH+x] = colorUCHAR(0, y, x); // writing output image   

    // test
    for (uint i=0; i<ls; ++i) { // ls<=7 ok, ls==8/9/10 crashes
      uint   adr  = WIDTH2*i+WIDTH*y+x; 
      border[adr] = (float3)(0);
    } 
}

Answer 1

边框 中存在对齐问题。在 cl_platform.h 中声明:

/* cl_float3 is identical in size, alignment and behavior to cl_float4. See section 6.1.5. */
typedef  cl_float4  cl_float3;

最好使用 cl_* 类型来避免此类问题，因此您可以按以下方式创建 border 缓冲区:

m_borderBuffer(getContext(), CL_MEM_READ_WRITE, (sizeof(cl_float3))*10*width*width), NULL),