如何防止 OpenCL 信号量中的死锁

Question

我正在遵循本教程，它一直都很棒，除了最后一个关于如何创建信号灯的示例对我不起作用。逻辑相当简单，但我无法弄清楚为什么这个内核会导致无限循环。

myKernel.cl

#pragma OPENCL EXTENSION cl_khr_global_int32_base_atomics : enable
void GetSemaphor(__global int * semaphor, __global int * data) {
   int occupied = atom_xchg(semaphor, 1);
   int realityCheck = 0;
   while(occupied == 1 && realityCheck++ < 100000)
        occupied = atom_xchg(semaphor, 1);
}
void ReleaseSemaphor(__global int * semaphor)
{
   int prevVal = atom_xchg(semaphor, 0);
}
__kernel void myKernel(__global int* data, __global int* semaphor)
{
    // semaphor[0] is set to 0 on the host.
    GetSemaphor(&semaphor[0], data);
    data[0]++;
    ReleaseSemaphor(&semaphor[0]);
}

这是与：

OpenCL 1.2

FULL_PROFILE

在具有 Quadro NVS

290 的 Quadro NVS 上

*cl_khr_global_int32_base_atomics cl_khr_global_int32_extended_atomics

Answer 1

您引用的教程是错误的，并且永远无法在 GPU 设备上运行。由于硬件要求。

阻止工作组内工作项的任何类型的同步机制都根本不起作用。由于阻塞状态会影响整个工作组，产生无限循环。

您只能在工作组大小为 1 的情况下执行此类操作。或跨工作组。