CUDA内核仅通过单个线程执行语句

如何在由单个线程执行的CUDA内核中编写语句。例如，如果我有以下的内核：CUDA内核仅通过单个线程执行语句

__global__ void Kernel(bool *d_over, bool *d_update_flag_threads, int no_nodes) 
{ 
    int tid = blockIdx.x*blockDim.x + threadIdx.x; 
    if(tid<no_nodes && d_update_flag_threads[tid]) 
    { 
    ... 
    *d_over=true; // writing a single memory location, only 1 thread should do? 
    ... 
    } 
}

在上述内核，“d_over”是单个布尔标志，而“d_update_flag_threads”是布尔阵列。

我通常做之前使用的线程块例如第一线索：

if(threadIdx.x==0)

，但它不能在这种情况下工作，我在这里有一个标志数组，只有线程用的产生密切相关标志“真“将执行if语句。该标志数组由另一个之前称为CUDA的内核设置，我事先没有任何关于它的知识。

总之，我需要类似于OpenMP中的“Single”构造。

来源

2012-06-05 usman

你格2维？ – geek

不，1维 – usman

为什么不简单地终止当前的'if'，为该赋值创建一个新的if（threadIdx.x == 0）'，然后用新的'if'恢复控制？ –

一种可能的方法是使用原子操作。如果每个块只需要一个线程来执行更新，则可以在共享内存中执行原子操作（对于计算能力> = 1.2），这通常比在全局内存中执行要快得多。

说，这个想法是如下：

int tid = blockIdx.x*blockDim.x + threadIdx.x; 

__shared__ int sFlag; 
// initialize flag 
if (threadIdx.x == 0) sFlag = 0; 
__syncthreads(); 

if(tid<no_nodes && d_update_flag_threads[tid]) 
{ 
    // safely update the flag 
    int singleFlag = atomicAdd(&sFlag, 1); 
    // custom single operation 
    if (singleFlag == 0) 
     *d_over=true; // writing a single memory location, only 1 thread will do it 
     ... 
}

这只是一个想法。我没有测试过它，但接近由单个线程执行的操作，而不是该块的第一个线程。

来源

2012-06-05 14:10:51 pQB

您可以使用atomicCAS（d_over，0，1），其中声明d_over或将其类型转换为int *。这将确保只有第一个看到d_over值为0（false）的线程才会更新它，而其他人不会。

来源

2012-06-05 23:47:13 Meghana

CUDA内核仅通过单个线程执行语句

回答

相关问题