0

我已经编写了一个 OpenCL 内核程序来将低通滤波器应用于图像。内核：

__kernel void applyLowPassFilter(__global int *image, __global int *rst,
                                 __local int *localMem) {
  int nCols = get_global_size(0); // width of image
  int nRows = get_global_size(1); // height of image

  int xg = get_global_id(0); // x index of global buffer
  int yg = get_global_id(1); // y index od global buffer

  int xl = get_local_id(0); // x index of local buffer

  localMem[xl] = image[yg * nCols + xg];
  barrier(CLK_LOCAL_MEM_FENCE);
  if (yg != 0) {
    rst[yg * nCols + xg] = (localMem[xl] + image[(yg - 1) * nCols + xg]) / 2;
  }
}

在内核代码中，我想访问每个工作组的本地内存并计算值。所以我将全局项目大小设置为 W*H（W：图像的宽度，H：图像的高度），将本地项目大小设置为 W*1，我期望组大小为 W 和组数这里的大小为 H。主机代码：

    size_t globalItemSize[2];
    size_t localItemSize[2];
    globalItemSize[0] = W;
    globalItemSize[1] = H;
    localItemSize[0] = W;
    localItemSize[1] = 1;
    // Set cl kernel arguments.
    ret = clSetKernelArg(clKernel, 0, sizeof(cl_mem), (void *)&imageObj);
    ret = clSetKernelArg(clKernel, 1, sizeof(cl_mem), (void *)&rstObj);
    ret = clSetKernelArg(clKernel, 2, sizeof(int) * localItemSize[0], NULL); // local mem

但是，代码不起作用，并不断给我一个零的结果图像。经过实验，我发现它只使用全局内存而不访问本地内存。我对访问本地内存的代码做错了吗？

4

1 回答 1

0

我通过监控OpenCL 返回错误代码发现了这一点。首先，我在通话CL_INVALID_KERNEL后收到 -48 错误代码。clSetKernelArg非常怀疑我的内核有问题。然后我删除了传递给内核以进行本地内存访问的第三个参数，并改用__local内核代码中的语句。此时，我收到 -51CL_INVALID_ARG_SIZE错误代码，提醒我使用该clinfo命令检查我的硬件的本地工作项编号限制。意识到本地项目大小的限制，我将localItemSize维度 0更改W为W/3. 然后它起作用了。

修改后的内核代码：

__kernel void applyLowPassFilter(__global int *image, __global int *rst) {
  int nCols = get_global_size(0); // width of image
  int nRows = get_global_size(1); // height of image

  int xg = get_global_id(0); // x index of global buffer
  int yg = get_global_id(1); // y index od global buffer

  int xl = get_local_id(0); // x index of local buffer

  __local int localMem[212]; // 1/3 of image width
  localMem[xl] = image[yg * nCols + xg];
  barrier(CLK_LOCAL_MEM_FENCE);
  if (yg != 0) {
    rst[yg * nCols + xg] = (localMem[xl] + image[(yg - 1) * nCols + xg]) / 2;
  }
}

主机代码中的参数配置：

    size_t globalItemSize[2];
    size_t localItemSize[2];
    globalItemSize[0] = W;
    globalItemSize[1] = H;
    localItemSize[0] = W / 3;
    localItemSize[1] = 1;
    // Set cl kernel arguments.
    ret = clSetKernelArg(clKernel, 0, sizeof(cl_mem), (void *)&imageObj);
    ret = clSetKernelArg(clKernel, 1, sizeof(cl_mem), (void *)&rstObj);
于 2020-04-29T21:57:42.573 回答