cudaStreamDestroy（）不同步/阻止？

我在Win7 x64機器上的Quadro NVS 295上使用CUDA 4.2。從CUDA C編程手冊我這樣說的：cudaStreamDestroy（）不同步/阻止？

」 ...流是通過調用cudaStreamDestroy釋放（）

for (int i = 0; i < 2; ++i) 
cudaStreamDestroy(stream[i]);

cudaStreamDestroy（）等待指定流中所有上面的命令。在破壞流並將控制權返回給主機線程之前完成。「

這是真的嗎？我寫了一個小的代碼，我做或多或少以下（我把唯一的僞代碼）：

//transfer input buffer to device 
cudaMemcpyToArrayAsync(... , stream[1]); 

//launch kernel 
my_kernel <<<dimGrid, dimBlock, 0, stream[1]>>> (...); 

//transfer from device to host 
cudaMemcpyAsync(.., cudaMemcpyDeviceToHost, stream[1]); 

//Destroy stream. In theory this should block the host until everything on the stream is completed! 
ret = cudaStreamDestroy(stream[1]);

有了這個例子，似乎cudaStreamDestroy（）調用立即返回到主機，即不等待爲cudaMemcpyAsync（）調用和其他strem指令完成。如果我把「cudaStreamSynchronize（stream [1]）;」電話befor銷燬流，一切順利，但速度較慢。那麼，我做錯了什麼？

非常感謝您的回覆！

來源

2012-06-11 ACRay

我不知道你在看什麼版本的文檔，但它不是我的一樣。我的CUDA 4.2文檔說：

銷燬並清理由stream指定的異步流。

如果該設備仍然在做工作的時候 cudaStreamDestroy（）被稱爲流流中，該功能將立即返回與流相關聯的資源一旦設備已完成流中的所有工作被自動釋放。

而且，根據我的經驗，這正是它的作用。驅動程序等待，直到流空並銷燬它。但cudaStreamDestroy不會阻止調用線程。

可以通過運行這個例子證實了這一點：

#include <stdio.h> 
#include <assert.h> 
#include <unistd.h> 

__global__ void kernel(int * inout, const int N) 
{ 
    int gid = threadIdx.x + blockIdx.x * blockDim.x; 
    int gstride = gridDim.x * blockDim.x; 

    for (; gid < N; gid+= gstride) inout[gid] *= 2; 
} 

#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); } 
inline void gpuAssert(cudaError_t code, char *file, int line, bool abort=true) 
{ 
    if (code != cudaSuccess) 
    { 
     fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line); 
     if (abort) exit(code); 
    } 
} 

int main(void) 
{ 
    const int N = 2<<20, sz = N * sizeof(int); 

    int * inputs, * outputs, * _inout; 

    gpuErrchk(cudaMallocHost((void **)&inputs, sz)); 
    gpuErrchk(cudaMallocHost((void **)&outputs, sz)); 
    gpuErrchk(cudaMalloc((void **)&_inout, sz)); 

    for(int i=0; i<N; i++) { inputs[i] = i; outputs[i] = 0; } 

    cudaStream_t stream[2]; 
    for (int i = 0; i < 2; i++) 
     gpuErrchk(cudaStreamCreate(&stream[i])); 

    gpuErrchk(cudaMemcpyAsync(_inout, inputs, sz, cudaMemcpyHostToDevice, stream[1])); 

    kernel<<<128, 128, 0, stream[1]>>>(_inout, N); 
    gpuErrchk(cudaPeekAtLastError()); 

    gpuErrchk(cudaMemcpyAsync(outputs, _inout, sz, cudaMemcpyDeviceToHost, stream[1])); 

    for(int i = 0; i < 2; i++) 
     gpuErrchk(cudaStreamDestroy(stream[i])); 

    sleep(1); // remove the sleep and see what happens.... 

    for(int i = 0; i < N; i++) 
     assert((2 * inputs[i]) == outputs[i]); 

    cudaDeviceReset(); 

    return 0; 
}

的無sleep()的代碼會失敗，因爲GPU還沒有完成，但有了它，assert會通過。請注意，sleep與cudaStreamDestroy調用之前使用顯式流同步基元的做法略有不同，即使結果相同。如果流在被銷燬時不是空的，那麼結果檢查將無法通過。

來源

2012-06-11 18:46:43 talonmies

非常感謝！我正在閱讀完全相同的文檔（4.2），但可能是不同的章節。我的意思是，在這一部分，它不是很清楚，我認爲這有點誤導。再次感謝您！ – ACRay

@ user1449129：如果這解決了你的問題，或許你可以如此友好[接受它]（http://meta.stackexchange.com/a/5235/163653）。 – talonmies

CUDA流僅僅是設備的任務執行隊列。所有接受流的函數只會將新任務添加到隊列中，而不會等待執行結果。 cudaStreamDestroy是一個特殊的任務，這意味着流需要被破壞，然後所有先前的設備任務完成。詞語

「cudaStreamDestroy（）等待給定的流破壞流和控制返回給主機線程之前完成在所有上面的命令」。

表示直到設備代碼完成才能銷燬流。

來源

2012-06-11 17:46:10 geek

謝謝！優秀的解釋！我喜歡這裏的nvidia文檔有點誤導！ – ACRay

cudaStreamDestroy（）不同步/阻止？

回答

相關問題