使用cudaEvent计时,也使用了cudaEventSynchronize(stopAll);计时结果为0是为什么呢

  • 5 replies
  • 2554 views
程序代码: [选择]
// Main.cu
#define  _CRT_SECURE_NO_WARNINGS
#define nFrame 2000

int main()
{
//译码程序计时
// struct timeval startAll,endAll;
// gettimeofday(&startAll, NULL);

cudaEvent_t startAll, stopAll;
float elapsedTime = 0.0;
cudaEventCreate(&startAll);
cudaEventCreate(&stopAll);

cudaEventRecord(startAll, 0);

genSource(src,mCol); //信源产生
rcmEncode(P,V, src, encOut,i,nFrame); //编码

//计算程序运行时间
cudaEventRecord(stopAll, 0);
cudaEventSynchronize(startAll);    //这句有没有结果都是0
cudaEventSynchronize(stopAll);

cudaEventElapsedTime(&elapsedTime, startAll, stopAll);
cudaEventDestroy(startAll);
cudaEventDestroy(stopAll);
printf("程序运行时间time = %f(ms)\n", elapsedTime);
printf("Processing Finish!!");
return 0;
}
// rcmEncode的实现

//kernel:编码,实质就是测量矩阵与权矢量相乘
__global__ void encode1(int *d_P,int* d_S,int* d_C)
{
unsigned int i = threadIdx.x + blockIdx.x * blockDim.x;
d_C[i] = -4 * d_S[d_P[i * wr + 0]] - 4 * d_S[d_P[i * wr + 1]] - 2 * d_S[d_P[i * wr + 2]] - d_S[d_P[i * wr + 3]]
     + d_S[d_P[i * wr + 4]] + 2 * d_S[d_P[i * wr + 5]] + 4 * d_S[d_P[i * wr + 6]] + 4 * d_S[d_P[i * wr + 7]];
}

void rcmEncode(int* P, int* V, int* src, int* encOut,int fi,int frame) {

//分配设备全局内存
int *d_S, *d_C,*d_P;

cudaMalloc((int**)&d_S, mCol*sizeof(int));//src
cudaMalloc((int**)&d_C, mea*sizeof(int));  //测量值
cudaMalloc((int**)&d_P, mea * wr * sizeof(int));//P
cudaMemset(d_C, 0, mea * sizeof(int));//初始化

//复制数据:从host --> device global,默认同步方式
cudaMemcpy(d_S, src, frameLen * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(d_P, P, mea * wr * sizeof(int), cudaMemcpyHostToDevice);

//分配线程和线程块:开启M个线程并行计算M个测量值,其中每个线程负责计算其中的一个测量值。
dim3 block(512);  //一个block的最大线程数为1024
dim3 grid((mea+block.x-1)/block.x);

//调用核函数
encode1 <<<grid, block >>>(d_P, d_S,d_C); //异步执行
//同步,编码完成后再进行数据的复制
cudaDeviceSynchronize();

//复制数据:从device --> host
cudaMemcpy(encOut, d_C, mea* sizeof(int), cudaMemcpyDeviceToHost);

//释放设备的全局内存(常量区不需要释放)
cudaFree(d_S);
cudaFree(d_C);
cudaFree(d_P);
}


程序代码: [选择]
// Main.cu
#define  _CRT_SECURE_NO_WARNINGS
#define nFrame 2000

int main()
{
//译码程序计时
// struct timeval startAll,endAll;
// gettimeofday(&startAll, NULL);

cudaEvent_t startAll, stopAll;
float elapsedTime = 0.0;
cudaEventCreate(&startAll);
cudaEventCreate(&stopAll);

cudaEventRecord(startAll, 0);

genSource(src,mCol); //信源产生
rcmEncode(P,V, src, encOut,i,nFrame); //编码

//计算程序运行时间
cudaEventRecord(stopAll, 0);
cudaEventSynchronize(startAll);    //这句有没有结果都是0
cudaEventSynchronize(stopAll);

cudaEventElapsedTime(&elapsedTime, startAll, stopAll);
cudaEventDestroy(startAll);
cudaEventDestroy(stopAll);
printf("程序运行时间time = %f(ms)\n", elapsedTime);
printf("Processing Finish!!");
return 0;
}
// rcmEncode的实现

//kernel:编码,实质就是测量矩阵与权矢量相乘
__global__ void encode1(int *d_P,int* d_S,int* d_C)
{
unsigned int i = threadIdx.x + blockIdx.x * blockDim.x;
d_C[i] = -4 * d_S[d_P[i * wr + 0]] - 4 * d_S[d_P[i * wr + 1]] - 2 * d_S[d_P[i * wr + 2]] - d_S[d_P[i * wr + 3]]
     + d_S[d_P[i * wr + 4]] + 2 * d_S[d_P[i * wr + 5]] + 4 * d_S[d_P[i * wr + 6]] + 4 * d_S[d_P[i * wr + 7]];
}

void rcmEncode(int* P, int* V, int* src, int* encOut,int fi,int frame) {

//分配设备全局内存
int *d_S, *d_C,*d_P;

cudaMalloc((int**)&d_S, mCol*sizeof(int));//src
cudaMalloc((int**)&d_C, mea*sizeof(int));  //测量值
cudaMalloc((int**)&d_P, mea * wr * sizeof(int));//P
cudaMemset(d_C, 0, mea * sizeof(int));//初始化

//复制数据:从host --> device global,默认同步方式
cudaMemcpy(d_S, src, frameLen * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(d_P, P, mea * wr * sizeof(int), cudaMemcpyHostToDevice);

//分配线程和线程块:开启M个线程并行计算M个测量值,其中每个线程负责计算其中的一个测量值。
dim3 block(512);  //一个block的最大线程数为1024
dim3 grid((mea+block.x-1)/block.x);

//调用核函数
encode1 <<<grid, block >>>(d_P, d_S,d_C); //异步执行
//同步,编码完成后再进行数据的复制
cudaDeviceSynchronize();

//复制数据:从device --> host
cudaMemcpy(encOut, d_C, mea* sizeof(int), cudaMemcpyDeviceToHost);

//释放设备的全局内存(常量区不需要释放)
cudaFree(d_S);
cudaFree(d_C);
cudaFree(d_P);
}


根据你的代码看,最大的可能是float elapsedTime根本就没有被写入,所以依然保持了初始值0.0f。

而没有写入的原因,最大的可能则是cudaEventElapsedTime()出错了,从而没有发挥正常的作用(返回2个Event的被record时刻的间隔)。

而该CUDA API函数出错的原因,看代码,最大的可能是之前的kernel或者CUDA函数挂掉了,从而导致后续持续的返回错误(CUDA Runtime API的一个特性这是)。

而具体哪里出错的(例如kernel还是api调用),因为你所有的kernel启动,同步时候的错误检测,API的返回值,都没有完全进行,我无法为你直接指出(但是你可以很容易的加上错误检测,自行定位)。

建议的解决方案:上一段的最后一句话。

根据你的代码看,最大的可能是float elapsedTime根本就没有被写入,所以依然保持了初始值0.0f。

而没有写入的原因,最大的可能则是cudaEventElapsedTime()出错了,从而没有发挥正常的作用(返回2个Event的被record时刻的间隔)。

而该CUDA API函数出错的原因,看代码,最大的可能是之前的kernel或者CUDA函数挂掉了,从而导致后续持续的返回错误(CUDA Runtime API的一个特性这是)。

而具体哪里出错的(例如kernel还是api调用),因为你所有的kernel启动,同步时候的错误检测,API的返回值,都没有完全进行,我无法为你直接指出(但是你可以很容易的加上错误检测,自行定位)。

建议的解决方案:上一段的最后一句话。

小提示:
直接使用了向上取整的block数量,例如你的:
dim3 grid((mea+block.x-1)/block.x);

然后kernel里面毫无任何的边界检测,这两个加起来很容易越界。

考虑到你的代码是节选的,mea不知道是什么具体值,我无法直接肉眼直接看出来。你自行检查错误。

小提示:
直接使用了向上取整的block数量,例如你的:
dim3 grid((mea+block.x-1)/block.x);

然后kernel里面毫无任何的边界检测,这两个加起来很容易越界。

考虑到你的代码是节选的,mea不知道是什么具体值,我无法直接肉眼直接看出来。你自行检查错误。
谢谢您!错误找到了,是在cudaMemcpy这个地方。
因为我的主函数是循环多次调用rcmEncode函数,因为cudaMalloc分配内存耗时还挺多,我看到有个博客说尽可能复用内存,减少内存回收次数。所以我就在第一个循环中分配内存,最后一次循环才释放内存。但是第二次循环时便出现code=1(cudaErrorInvalidValue) "cudaMemcpy(d_S, src, mCol * sizeof(int), cudaMemcpyHostToDevice)" 这个错误,请问设备内存不能这样复用吗?第二次为同一个变量所在的设备内存进行复制不会覆盖前一次循环所复制的值吗

谢谢您!错误找到了,是在cudaMemcpy这个地方。
因为我的主函数是循环多次调用rcmEncode函数,因为cudaMalloc分配内存耗时还挺多,我看到有个博客说尽可能复用内存,减少内存回收次数。所以我就在第一个循环中分配内存,最后一次循环才释放内存。但是第二次循环时便出现code=1(cudaErrorInvalidValue) "cudaMemcpy(d_S, src, mCol * sizeof(int), cudaMemcpyHostToDevice)" 这个错误,请问设备内存不能这样复用吗?第二次为同一个变量所在的设备内存进行复制不会覆盖前一次循环所复制的值吗
再次谢谢!我把int *d_S, *d_C,*d_P;这几个变量设置为全局变量之后可以了

谢谢您!错误找到了,是在cudaMemcpy这个地方。
因为我的主函数是循环多次调用rcmEncode函数,因为cudaMalloc分配内存耗时还挺多,我看到有个博客说尽可能复用内存,减少内存回收次数。所以我就在第一个循环中分配内存,最后一次循环才释放内存。但是第二次循环时便出现code=1(cudaErrorInvalidValue) "cudaMemcpy(d_S, src, mCol * sizeof(int), cudaMemcpyHostToDevice)" 这个错误,请问设备内存不能这样复用吗?第二次为同一个变量所在的设备内存进行复制不会覆盖前一次循环所复制的值吗

如果cudaMemcpy提示invalid value的话,你得好好检查你的参数了,而不是怀疑cudaMemcpy()能否调用两次(因为这显然的)。

自行检查你的d_S, src, mCol等值是否正常。特别是d_S的分配是否成功。