用計數器替換blockId

Question

最初我在我的代碼中使用了blockIdx.x，但我想刪除它，而是有一個全局值，並在我的塊中使用它而不是blockidx.x。由於我的代碼太大，當我以大輸入尺寸運行代碼時，它會掛起，我認爲它可以提供幫助。我自動增加計數器，但是當我運行代碼時，它會掛起。任何人都可以看看我的代碼，看看我是否做錯了什麼？

__device__ int counter = 0; 

__global__ void kernel(int * ginput, int * goutput) 
{ 
    const int tid = threadIdx.x; 
    const int id = threadIdx.x + blockIdx.x * blockDim.x; 
    in myval = ginput[id]; 

    if (tid == 0) { 
    atomicAdd(&counter, 1); 
    } 

    __syncthreads(); 
    if (counter == 0) { 
    goutput[tid] = ...; 
    } 
    if (counter > 0) { 
    ... 
    } 

} 

如果我使用的，而不是在我的代碼計數器blockIdx.x它的工作原理，但我只是想反

Answer 1

來取代它。如果你想counter更換的blockIdx.x您的使用情況（例如，你希望每一塊有它從counter讀取唯一值），那麼這樣的事情應該工作：

__device__ int counter = 0; 

__global__ void kernel(int * ginput, int * goutput) 
{ 
    const int tid = threadIdx.x; 
    const int id = threadIdx.x + blockIdx.x * blockDim.x; 
    __shared__ int my_block_id; 


    if (tid == 0) { 
    my_block_id = atomicAdd(&counter, 1); 
    } 

    __syncthreads(); 
    if (my_block_id == 0) { 
    goutput[tid] = ...; 
    } 
    if (my_block_id > 0) { 
    ... 
    } 

} 

你的方法將是麻煩的，因爲如果你做這樣的事情：

if (counter > 5) .... 

你可能讀從全局內存的counter一個新的更新值，以及任意數量的塊可能有更新的價值，所以其行爲將是不可預測的。