cuda使用循環緩衝區的核心外實現

我試圖在GPU內存和CPU內存之間進行外部核心。例如，我有1GB的數據塊，我需要按順序處理1000個這樣的塊，每個塊都通過內核啓動完成。假設處理必須一個接一個完成，因爲第n個內核啓動需要使用第（n-1）個內核產生的結果，該內核存儲在第（n-1）個塊中，第一次內核啓動。所以我正考慮在GPU上使用循環緩衝區來存儲最近的5個塊，並使用事件在數據流和任務流之間進行同步。數據流準備數據，任務流啓動內核。代碼如下所示。cuda使用循環緩衝區的核心外實現

const int N_CBUF = 5, N_TASK = 1000; 

// Each pointer points to a data block of 1GB 
float* d_cir_buf[N_CBUF]; 
float* h_data_blocks[N_TASK]; 

// The data stream for transfering data from host to device. 
// The task stream for launching kernels to process the data. 
cudaStream_t s_data, s_task; 

// The data events for the completion of each data transfer. 
// The task events for the completion of each kernel execution. 
cudaEvent_t e_data[N_TASK], e_task[N_TASK]; 

// ... code for creating the streams and events. 

for (int i = 0; i < N_TASK; i++) { 
    // Data transfer should not overwritten the data needed by the kernels. 
    if (i >= N_CBUF) { 
    cudaStreamWaitEvent(s_data, e_task[i-N_CBUF+1]); 
    } 
    cudaMemcpyAsync(d_cir_buf[i % N_CBUF], h_data_blocks[i], ..., cudaMemcpyHostToDevice, s_data); 
    cudaEventRecord(e_data[i], s_data); 

    cudaStreamWaitEvent(s_task, e_data[i]); 

    // Pass the current and the last data block to the kernel. 
    my_kernel<<<..., s_task>>>(d_cir_buf[i % N_CBUF], 
    i == 0 ? 0 : d_cir_buf[(i+N_CBUF-1)%N_CBUF]); 
    cudaEventRecord(e_task[i], s_task); 
}

我想知道這是否是一個有效的想法，或者是否有什麼完全錯誤的？另外，CUDA編程指南提到，如果從兩個不同的主機內存地址到同一個設備地址有memcpy，那麼將不會有併發執行，這對我而言是否有用？特別是，如果d_cir_buf的內存被分配爲一個整體的大塊，然後分成5塊，那麼它會被視爲「設備中的相同內存地址」，從而導致併發性失敗？另外，在我的情況下，第（n + 5）個數據傳輸將與第n個數據傳輸一樣到達相同的地址，但是，考慮到需要同步，將不會有兩個這樣的傳輸同時執行。那麼這是好的嗎？

來源

2014-02-28 shaoyl85

我有你的問題是最適合雙緩衝的感覺：

兩股氣流在流1
上傳數據1
上傳數據2的數據1
運行的內核
在stream2的data2上運行內核

...等

核心在流2可以與數據傳輸在strezm 1重疊，反之亦然

來源

2014-03-01 22:48:06 Sigismondo

謝謝，雙緩衝是我的程序的特殊情況設置'N_CBUF'到2的時候，但我需要至少3個，因爲每個內核啓動需要使用2個連續的塊。我的問題是，這種實現雙/三/ N-元組緩衝的方式是正確的嗎？或者是否有任何問題會禁用併發？ – shaoyl85

據我所知，雙緩衝是唯一的選擇，以防一次運行一個內核。重疊的計算和內存傳輸是目標，所以你肯定可以讓這兩個內核獲得兩個以上的元素/塊。這也是雙緩衝。唯一非常重要的是，你不要觸摸正在運行的內核使用的隊列塊，同時以[這種方式]準備下一個數據的數據（http://docs.nvidia.com/cuda/cuda- – Sigismondo

我明白了，你是對的，我並不打算在這種情況下讓不同的內核在同一時間運行，但是我錯過了同步以防止這種情況發生發生。 – shaoyl85

cuda使用循環緩衝區的核心外實現

回答

相關問題