使用CUDA可以在網格中選擇一些塊嗎？

我的源代碼是對圖像的遞歸作業。但在處理中，許多區塊滿足約8倍的結束條件。只有幾個塊導致執行循環超過16次。所以我想跳過滿足執行結束條件的塊。

這是可能的嗎？

__global__ main(){ 
/* previous */ 
int *blockMap; 
cudaMalloc((void**)&blockMap, sizeof(int) * nXBlockNum * nYBlockNum); 
cudaMemset((void**)&blockMap, 0, sizeof(int) * nXBlockNum * nYBlockNum); 

kernel<<<nblocks, nthreads>>>(inputimage, outputbuffer, blockmap); 
/* after */} 

__global__ 
kernel(byte* inputeimage, byte* outputbuffer, int* blockmap) { 
    __shared__ int *skipFlag; 


    if((blockDim.x * threadIdx.y + threadIdx.x) == 0) 
    { 
     *skipFlag = g_bMap[blockIdx.y * gridDim.x + blockIdx.x]; 
    } 

    if(*skipFlag == 0) 
    { 
       /* recursive job */ 
    } 
}

來源

2013-03-15 user2174291

你是什麼意思？如果你有'if'條件，你可以繼續只運行活動塊。即使使用塊內的線程，也可以這樣做。 – 2013-03-15 14:35:00

簡單的方法是爲所有線程大小聲明標誌。但我想使用具有塊大小的標誌變量。所以我宣佈標誌變量在全局內存中，比使用塊索引，以便它不運行。但在運行時間，它似乎是死鎖狀態。 – user2174291 2013-03-15 14:47:34

我不太瞭解你的問題，但這種方式似乎是正確的。每個塊在它們的共享內存中都有一個獨特的skipFlag，而那些得到true的塊不會執行其餘的代碼。

而且，也許兩個ifs之間的__syncthreads()將是一個好主意。

來源

2013-03-15 16:06:07 machlas

是的，你可以做到這一點，但你展示的內核代碼並不完全是這樣做的。假設你想要一個整數標誌每個塊，則代碼應該是這個樣子：

__global__ 
kernel(byte* inputeimage, byte* outputbuffer, int* blockmap) { 
    __shared__ int skipFlag; 


    if (threadIdx.x == 0) 
    { 
     skipFlag = g_bMap[blockIdx.x]; 
    } 
    __syncthreads(); 

    if(skipFlag == 0) 
    { 
       /* recursive job */ 
    } 
}

這裏的第一個線程中的每個塊加載該特定塊的全球並將其存儲到共享內存整數標誌變量。在塊同步之後，每個線程都可以讀取該值並相應地對其執行操作。

來源

2013-03-15 16:16:08 talonmies

謝謝。我會嘗試這一個。謝謝alob – user2174291 2013-03-15 16:28:15

使用CUDA可以在網格中選擇一些塊嗎？

回答

相關問題