OpenCL for 循环执行模型

Question

我目前正在学习 OpenCL 并遇到了这个代码片段：

int gti = get_global_id(0);
int ti = get_local_id(0);

int n = get_global_size(0);
int nt = get_local_size(0);
int nb = n/nt;

for(int jb=0; jb < nb; jb++) { /* Foreach block ... */
      pblock[ti] = pos_old[jb*nt+ti]; /* Cache ONE particle position */
      barrier(CLK_LOCAL_MEM_FENCE); /* Wait for others in the work-group */

      for(int j=0; j<nt; j++) { /* For ALL cached particle positions ... */
         float4 p2 = pblock[j]; /* Read a cached particle position */
         float4 d = p2 - p;
         float invr = rsqrt(d.x*d.x + d.y*d.y + d.z*d.z + eps);
         float f = p2.w*invr*invr*invr;
         a += f*d; /* Accumulate acceleration */
      }

      barrier(CLK_LOCAL_MEM_FENCE); /* Wait for others in work-group */
}

关于代码的背景信息：这是 NBody 模拟程序中 OpenCL 内核的一部分。可以在此处找到完整的代码和教程。

这是我的问题（主要与 for 循环有关）：

1. 在 OpenCL 中 for 循环究竟是如何执行的？我知道所有工作项都运行相同的代码，并且工作组中的工作项尝试并行执行。因此，如果我在 OpenCL 中运行 for 循环，这是否意味着所有工作项都运行相同的循环，还是以某种方式将循环划分为跨多个工作项运行，每个工作项执行循环的一部分（即工作项1 处理索引 0 ~ 9，项目 2 处理索引 10 ~ 19 等）。

2. 在这段代码片段中，外循环和内循环如何执行？OpenCL 是否知道外循环在所有工作组之间分配工作，而内循环试图在每个工作组内的工作项之间分配工作？

3. 如果内部循环在工作项之间划分（意味着 for 循环中的代码并行执行，或者至少尝试并行执行），那么最后的加法如何工作？它本质上是在做 a = a + f*d，根据我对流水线处理器的理解，这必须按顺序执行。

我希望我的问题足够清楚，我感谢任何输入。

Answer 1

异构编程适用于工作分配模型，这意味着线程得到其应有的部分来工作并启动。

1.1）如您所知，线程是在工作组（或线程块）中组织的，在您的情况下，工作组（或线程块）中的每个线程将数据从全局内存带到本地内存。

for(int jb=0; jb < nb; jb++) { /* Foreach block ... */
      pblock[ti] = pos_old[jb*nt+ti];

//I assume pblock is local memory

1.2）现在线程块中的所有线程都在本地存储中拥有它们所需的数据（因此无需再访问全局内存）

1.3）现在是处理，如果你仔细观察发生处理的 for 循环

for(int j=0; j<nt; j++) {

其运行线程块总数。因此，这个循环片段设计确保所有线程处理单独的数据元素。

1）for循环就像OpenCL的另一个C语句一样，所有线程都会按原样执行它，这取决于你如何划分它。OpenCL 不会在内部为您的循环执行任何操作（如第 1.1 点）。

2）OpenCL对你的代码一无所知，它不知道你如何划分循环。

3）与语句相同：1内循环不划分在线程之间，所有线程都会按原样执行，唯一的事情是它们会指向它们想要处理的数据。

我想你的这种困惑是因为你在对线程块和本地内存有很多了解之前就跳入了代码。我建议您查看此代码的初始版本，其中根本没有使用本地内存。