OpenCL - 重用全局内存

Question

这是一个看似基本的问题，我通过大量的反复试验无法解决。我有一个内核，它使用两个全局 r/w 缓冲区和一个本地 - 它从第一个缓冲区获取输入，使用第二个缓冲区对其进行伪排序以进行临时存储，并最终将其复制回第一个缓冲区一定的顺序。 （剥离）代码如下：

struct PACKET_POINTER {
       int packetIndex;
       int currentCell;
};

#define RPC_DIV_BUCKET 100
__kernel void PseudoSort(__global struct PACKET_POINTER * in,__global struct PACKET_POINTER * out, __local struct PACKET_POINTER * aux) {
  int i = get_local_id(0);
  int wg = get_local_size(0);
  int gid = get_global_id(0);
  int offset = get_group_id(0) * wg;

  aux[i] = in[i+offset];
  barrier(CLK_LOCAL_MEM_FENCE);
  //-----
  //Irrelevant code block here
  //----- 
  out[(gid%1024)*RPC_DIV_BUCKET + (gid/1024)] = aux[i];
}

在父 C 程序中检索“out”缓冲区的内容没有问题。但是，当我将以下几行添加到内核时：

    barrier(CLK_GLOBAL_MEM_FENCE);
    in[gid] = out[gid];

并尝试读取“in”缓冲区，它在第一次执行时主要出现垃圾值，但如果 .exe 在没有修改的情况下第二次运行，它将获得预期的数据。我在内核调用和缓冲区读取之间有一个 clFinish(commands) 调用，因此它应该在任何读取尝试之前运行完成。我在这里遗漏了一些明显的东西？提前感谢您的帮助 - 如果我在那之前遇到它会发布解决方案。

Answer 1

CLK_GLOBAL_MEM_FENCE 仅在工作组内同步。没有办法设置可以在所有工作组之间同步的屏障（例如，它只在具有相同 group_id 的线程之间同步）。

你在那里有一个竞争条件。例如，当 global_id 为 1 时，写入会进入 out[100]。然后该特定线程从 out[1] 读取并写入 in[1]。但是 out[1] 仅在 global_id 1024 处写入。几乎可以肯定它在不同的工作组中。所以你会读到垃圾，因为第一个工作组将在 out[1] 被写入之前完成。