如何防止OpenCL信號中的死鎖？

我正在關注this教程，除了關於如何創建信號燈的最後一個例子，我一直都很贊。邏輯非常簡單，但我無法弄清楚爲什麼這個內核導致無限循環。如何防止OpenCL信號中的死鎖？

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具有

* cl_khr_global_int32_base_atomics CL _khr_global_int32_extended_atomics

來源

2013-09-27 user1873073

你所反饋的那個教程是錯誤的，並且不會在GPU設備上工作。由於硬件架構。

任何阻止工作組內工作項的同步機制都不起作用。由於阻塞狀態將影響整個工作組，產生無限循環。

您只能在工作組大小爲1或者跨工作組的情況下完成這些工作。

來源

2013-09-28 07:58:19 DarkZeros

很好的答案，謝謝！ – user1873073

同意。問題是第一個工作項獲得信號量，然後其餘工作項都不能繼續執行，除了在大多數體系結構中，許多工作項都綁在一起成爲warp或wavefront，並且必須以鎖步，所以通過阻止其他工作項目，您會使整個warp/wavefront發生死鎖，從而導致工作組發生死鎖，從而導致內核死鎖。 – Dithermaster

如何防止OpenCL信號中的死鎖？

回答

相關問題