1

我有一个 OpenCL 内核，其中每个工作组在本地内存中生成一个结果向量。然后我需要将所有这些结果汇总到全局内存中，以便稍后检索到主机。
为了测试这一点，我创建了以下内核代码：

//1st thread in each workgroup initializes local buffer
if(get_local_id(0) == 0){
    for(i=0; i<HYD_DIM; i++){
        pressure_Local[i] = (float2){1.0f, 0.0f};
    }
}

//wait for all workgroups to finish accessing any memory
barrier(CLK_GLOBAL_MEM_FENCE | CLK_LOCAL_MEM_FENCE);

/// sum all the results into global storage
for(i=0; i<get_num_groups(0); i++){

    //1st thread in each workgroup writes the group's local buffer to global memory
    if(i == get_group_id(0) && get_local_id(0) == 0){
        for(j=0; j<HYD_DIM; j++){
            pressure_Global[j] += pressure_Local[j];
            // barrier(CLK_GLOBAL_MEM_FENCE);
        }
    }

    //flush global memory buffers:
    barrier(CLK_GLOBAL_MEM_FENCE);
}

本质上，我期望全局内存中向量的所有元素都等于工作组的数量（在我的例子中是 128 个）。实际上，它们通常在 60 到 70 之间变化，并且结果因运行而异。
有人可以告诉我我错过了什么，或者如何正确地做到这一点？

4

1 回答 1

2

您无法使用 opencl 在不同的工作组之间进行同步。CLK_GLOBAL_MEM_FENCE 不能那样工作。它只保证将保持内存操作（由工作组访问）的顺序。请参阅OCL 1.2 规范中的“6.12.8 同步函数”部分。

我会通过为每个工作组使用不同的全局内存块来解决您的问题。你将数据写入全局，你的内核就完成了。然后，如果您想将数据减少到单个块，您可以让另一个内核从全局读取数据，并将其与其他结果块合并。您可以根据需要进行任意多层合并，但最终合并必须由单个工作组完成。

Search around for gpu/opencl reduction algorithms. Here's a decent one to start with. Case Study: Simple Reductions

于 2012-09-21T17:42:38.503 回答