同步在CUDA

Question

我讀的CUDA有關同步CUDA參考手冊，但我不知道很清楚。例如爲什麼我們使用cudaDeviceSynchronize()或__syncthreads()？如果不使用它們會發生什麼情況，程序無法正常工作？什麼cudaMemcpy和行動cudaMemcpyAsync區別？
你能展示一個顯示這種差異的例子嗎？

Answer 1

cudaDeviceSychronize()在主機代碼使用（即在CPU上運行）當期望的是CPU活動等待任何未決GPU活動的完成。在許多情況下，這是沒有必要明確地做到這一點，作爲頒發給單流GPU操作是自動序列，像cudaMemcpy()某些其他操作都內置了固有的閉鎖裝置同步。但出於其他目的，例如調試代碼，強制設備完成任何未完成的活動可能會很方便。

__syncthreads()用於設備代碼（即在GPU上運行），並且在具有獨立並行操作的代碼中（例如逐個添加兩個向量）可能根本不需要。但是，常用的一個例子是運行在共享內存之外的算法。在這種情況下，它是經常需要從全局內存值加載到共享內存中，我們希望在threadblock每個線程有機會加載它是適當的共享存儲位置（S），發生任何實際的處理之前。在這種情況下，我們要使用__syncthreads()發生的處理之前，確保共享內存完全填充。這只是一個例子。 __syncthreads()可能用於任何需要線程塊內的任何時間同步。它不允許塊之間的同步。

cudaMemcpy和cudaMemcpyAsync之間的區別在於該調用的非異步版本只能發送到流0，並會阻止調用CPU線程，直到複製完成。異步版本可以選擇使用流參數，並在複製完成之前立即將控制權返回給調用線程。異步版本通常在我們想要擁有asynchronous concurrent execution的情況下找到用法。

如果您有關於CUDA編程的基本問題，建議您採用一些webinars available。

Answer 2

此外，__syncthreads()當你在你的代碼中的一些有條件的路徑，然後你想運行取決於幾個數組元素的操作變得很必要的。 考慮下面的例子：

int n = threadIdx.x; 

if(myarray[n] > 0) 
{ 
    myarray[n] = - myarray[n]; 
} 
double y = myarray[n] + myarray[n+1]; // Not all threads reaches here at the same time 

在上述例子中，不是所有的線程將具有相同的執行順序。根據if條件，某些線程需要更長的時間。在考慮示例的最後一行時，您需要確保所有線程都完全正確完成了if條件並更新了myarray。如果不是這種情況，y可能會使用一些更新和未更新的值。 在這種情況下，它成爲一個必須要評估y來克服這個問題前添加__syncthreads()：

if(myarray[n] > 0) 
{ 
    myarray[n] = - myarray[n]; 
} 
__syncthreads(); // All threads will wait till they come to this point 
// We are now quite confident that all array values are updated. 
double y = myarray[n] + myarray[n+1];