1

Win7x64マシンのQuadroNVS295でCUDA4.2を使用しています。CUDA Cプログラミングマニュアルから私はこれを読みました：

"...ストリームは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]）;」を入れたら ストリームを破棄する前に呼び出してください。すべてがうまくいきますが、遅くなります。だから、私が間違っているのは何ですか？

ご回答ありがとうございます！

4

2 に答える 2

3

あなたが見ているドキュメントのバージョンはわかりませんが、私のものと同じではありません。私の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 が完成していないため、コードがないと失敗しますが、GPU があれば合格assertします。結果が同じであっても、呼び出しのsleep前に明示的なストリーム同期プリミティブを使用するのとは微妙に異なることを行っていることに注意してください。cudaStreamDestroyストリームが破棄されたときにストリームが空でなかった場合、結果チェックに合格することはありません。

于 2012-06-11T18:46:43.563 に答える
2

CUDA ストリームは、デバイス タスクの単なる実行キューです。ストリームを受け入れるすべての関数は、実行結果を待たずに新しいタスクをキューに追加するだけです。cudaStreamDestroyこれは、ストリームを破棄する必要があることを意味する特別なタスクであり、その後、以前のすべてのデバイス タスクが完了します。言葉

「cudaStreamDestroy() は、指定されたストリーム内の先行するすべてのコマンドが完了するのを待ってから、ストリームを破棄して制御をホスト スレッドに戻します。」

これは、デバイス コードが完成するまでストリームを破棄できないことを意味します。

于 2012-06-11T17:46:10.690 に答える