以下代码用于测试 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 ]
不知道为什么...