1

次のコードは、cudaMemcpyAsyncの同期動作をテストするためのものです。

#include <iostream>
#include <sys/time.h>

#define N 100000000

using namespace std;


int diff_ms(struct timeval t1, struct timeval t2) 
{
    return (((t1.tv_sec - t2.tv_sec) * 1000000) +
            (t1.tv_usec - t2.tv_usec))/1000;
}

double sumall(double *v, int n)
{
    double s=0;
    for (int i=0; i<n; i++) s+=v[i];
    return s;
}


int main()
{
    int i;

    cudaStream_t strm;
    cudaStreamCreate(&strm);

    double *h0;
    double *h1;
    cudaMallocHost(&h0,N*sizeof(double));
    cudaMallocHost(&h1,N*sizeof(double));

    for (i=0; i<N; i++) h0[i]=99./N;
    double *d; 
    cudaMalloc(&d,N*sizeof(double));

    struct timeval t1, t2; gettimeofday(&t1,NULL);
    cudaMemcpyAsync(d,h0,N*sizeof(double),cudaMemcpyHostToDevice,strm);
    gettimeofday(&t2, NULL); printf("cuda H->D %d takes: %d ms\n",i, diff_ms(t2, t1)); gettimeofday(&t1, NULL);
    cudaMemcpyAsync(h1,d,N*sizeof(double),cudaMemcpyDeviceToHost,strm);
    gettimeofday(&t2, NULL); printf("cuda D->H %d takes: %d ms\n",i, diff_ms(t2, t1)); gettimeofday(&t1, NULL);

    cout<<"sum h0: "<<sumall(h0,N)<<endl;
    cout<<"sum h1: "<<sumall(h1,N)<<endl;


    cudaStreamDestroy(strm);
    cudaFree(d);
    cudaFreeHost(h0);
    cudaFreeHost(h1);

    return 0;
}

h0 / h1のプリントアウトは、cudaMemcpyAsyncがホストと同期していることを示しています

sum h0: 99
sum h1: 99

ただし、cudaMemcpyAsync呼び出しを含む時間差は、それらがホストと同期されていないことを示しています

cuda H->D 100000000 takes: 0 ms
cuda D->H 100000000 takes: 0 ms

これはcuda-profilingの結果ではサポートされていないためです。

method=[ memcpyHtoDasync ] gputime=[ 154896.734 ] cputime=[ 17.000 ] 
method=[ memcpyDtoHasync ] gputime=[ 141175.578 ] cputime=[ 6.000 ]

理由がわからない...

4

1 に答える 1

5

ここでは (少なくとも) 2 つのことが行われています。

あなたの最初の観察はそれです：

sum h0: 99
sum h1: 99

同じストリームに対して発行されたCUDA 呼び出しは、順番に実行されます。CUDA 呼び出しを互いにオーバーラップさせたい場合は、別のストリームに発行する必要があります。cuda memcpy をデバイスに発行し、デバイスから同じストリームで発行しているため、それらは順番に実行されます。2 番目のものは、最初のものが完了するまで開始されません (両方ともすぐにキューに入れられますが)。したがって、データは無傷で (最初の cudaMemcpy の後)、両方の配列が適切な合計を生成することがわかります。

あなたの残りの観察も互いに一致しています。あなたは次のことを報告しています:

cuda H->D 100000000 takes: 0 ms
cuda D->H 100000000 takes: 0 ms

これは、これらの非同期呼び出しの両方がすぐにホスト スレッドに制御を返し、呼び出しがキューに入れられて、ホストの実行とは非同期に実行されるためです。その後、呼び出しはさらにホストの実行と並行して進行します。制御はすぐにホストに返され、ホストベースのタイミング メソッドを使用して操作のタイミングを合わせているため、時間がかからないように見えます。

もちろん、実際には時間がゼロというわけではなく、プロファイラーの結果はそれを示しています。GPU は CPU に対して非同期で実行されているため (この場合は cudaMemcpyAsync を含む)、プロファイラーは cudaMemcpy 操作にかかる実際の「リアルタイム」時間を表示gputimeし、CPU の「見かけの時間」、つまり時間も報告します。として報告された操作を起動するために必要な CPU cputime。cputime は gputime に比べて非常に小さいことに注意してください。つまり、ほとんど瞬間的であるため、ホスト ベースのタイミング メソッドではゼロ時間が報告されます。しかし、実際には完了するまでに時間が 0 というわけではなく、プロファイラーはそれを報告します。

cudaEvent タイミング メソッドを使用した場合、もちろん異なる結果が表示され、プロファイラーのgputime結果により近くなります。

于 2012-12-22T11:54:40.603 に答える