CUDA：從內核調用__device__函數

Question

我有一個內核調用if語句中的設備函數函數。代碼如下：

__device__ void SetValues(int *ptr,int id) 
{ 
    if(ptr[threadIdx.x]==id) //question related to here 
      ptr[threadIdx.x]++; 
} 

__global__ void Kernel(int *ptr) 
{ 
    if(threadIdx.x<2) 
     SetValues(ptr,threadIdx.x); 
} 

在內核線程0-1中同時調用SetValues。那之後會發生什麼？我的意思是現在有2個併發呼叫SetValues。每個函數調用是否都是串行執行的？所以他們的行爲就像2個內核函數調用一樣？

Answer 1

CUDA默認內聯了所有函數（雖然Fermi和更新的架構也支持帶有函數指針和實函數調用的正確ABI）。所以，你的示例代碼被編譯到這樣的事情

__global__ void Kernel(int *ptr) 
{ 
    if(threadIdx.x<2) 
     if(ptr[threadIdx.x]==threadIdx.x) 
      ptr[threadIdx.x]++; 
} 

執行並行發生，就像正常的代碼。如果你將一個內存競爭設計成一個函數，那麼沒有序列化機制可以爲你節省。