CUDA C - 使用时钟（）和 cudaEvent 的 CPU 和 GPU 执行时间，是否正确？

Question

我写了一个程序来添加两个二维数组来检查 CPU 和 GPU 的性能。 我使用 clock() 函数来测量 CPU 执行和 cudaEvent 来测量 GPU 中的内核执行时间。由于我是在 Udacity 下学习 CUDA，所以我尝试在他们的服务器上执行程序，发现结果为，

 Output:
 GPU: 0.001984 ms
 CPU : 30.000000 ms

现在回到我真正的问题，我发现这些结果在 GPU 上的速度非常快，现在我有点怀疑这些结果是否准确，或者我的程序是否有任何错误？

这是我的程序：

 #include "stdio.h"
 #include<time.h>
 #define COLUMNS 900
 #define ROWS 900
 long a[ROWS][COLUMNS], b[ROWS][COLUMNS], c[ROWS][COLUMNS],d[ROWS][COLUMNS];
__global__ void add(long *a, long *b, long *c,long *d)
{
 int x = blockIdx.x;
 int y = blockIdx.y;
 int i = (COLUMNS*y) + x;
 c[i] = a[i] + b[i];
 a[i]=d[i];
}

int main()
{
  long *dev_a, *dev_b, *dev_c,*dev_d;
  float ms;
  clock_t startc, end;
  double cpu_time_used;
  cudaEvent_t start,stop;


 cudaMalloc((void **) &dev_a, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_b, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_c, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_d, ROWS*COLUMNS*sizeof(int));

 startc = clock();
 for (long y = 0; y < ROWS; y++) // Fill Arrays
 for (long x = 0; x < COLUMNS; x++)
 {
     a[y][x] = x;
     b[y][x] = y;
     d[y][x]=rand()%4;
     c[y][x]=a[y][x]+b[y][x];
 }
 end = clock();

cpu_time_used = ((double) (end - startc)) / CLOCKS_PER_SEC;
cpu_time_used*=1000;


cudaMemcpy(dev_a, a, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);
cudaMemcpy(dev_b, b, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);
cudaMemcpy(dev_d, d, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);


cudaEventCreate(&start);
cudaEventCreate(&stop);
cudaEventRecord(start, 0); 
cudaEventRecord(stop, 0);


add<<<dim3(1024,1024),dim3(128,128)>>>(dev_a, dev_b, dev_c,dev_d);

cudaEventSynchronize(stop);
cudaEventElapsedTime(&ms, start, stop);
cudaMemcpy(c, dev_c, ROWS*COLUMNS*sizeof(int),cudaMemcpyDeviceToHost);
cudaEventDestroy(start);
cudaEventDestroy(stop);




printf("GPU: %f ms",ms);
printf("\n CPU : %f ms",cpu_time_used);

 return 0;
}

感谢大家对我的查询提供的答案，这是我对代码所做的更改和更新的结果，

更新代码：

#include "stdio.h"
#include <time.h>
#include <sys/time.h>
#include <unistd.h>
#define COLUMNS 500
#define ROWS 500
long a[ROWS][COLUMNS], b[ROWS][COLUMNS], c[ROWS][COLUMNS],d[ROWS][COLUMNS];



__global__ void add(long *a, long *b, long *c,long *d)
{
 int x = blockIdx.x;
 int y = blockIdx.y;
 int i = (COLUMNS*y) + x;
 c[i] = a[i] + b[i];
 a[i]=d[i];
}
int main()
{
 long *dev_a, *dev_b, *dev_c,*dev_d;
 struct timeval startc, end;
 float ms;
 long mtime, seconds, useconds;
 //   clock_t startc, end;
 //  double cpu_time_used;
 long ns;
 cudaEvent_t start,stop;


 cudaMalloc((void **) &dev_a, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_b, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_c, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_d, ROWS*COLUMNS*sizeof(int));

 gettimeofday(&startc, NULL);
 for (long y = 0; y < ROWS; y++) // Fill Arrays
 for (long x = 0; x < COLUMNS; x++)
 {
  a[y][x] = x;
  b[y][x] = y;
  d[y][x]=rand()%4;
  c[y][x]=a[y][x]+b[y][x];
 }
  gettimeofday(&end, NULL);

 seconds  = end.tv_sec  - startc.tv_sec;
 useconds = end.tv_usec - startc.tv_usec;
 mtime = ((seconds) * 1000 + useconds/1000.0) + 0.5;


for (long y = ROWS-1; y < ROWS; y++) // Output Arrays
 {
 for (long x = COLUMNS-1; x < COLUMNS; x++)
 {
    // printf("\n[%ld][%ld]=%ld ",y,x,c[y][x]);
   //   printf("[%d][%d]=%d ",y,x,d[y][x]);
 }
 printf("\n");
 }



cudaMemcpy(dev_a, a, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);
cudaMemcpy(dev_b, b, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);
cudaMemcpy(dev_d, d, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);


cudaEventCreate(&start);
cudaEventCreate(&stop);
cudaEventRecord(start, 0); 



add<<<dim3(1024,1024),dim3(128,128)>>>(dev_a, dev_b, dev_c,dev_d);

cudaThreadSynchronize();
cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
cudaEventElapsedTime(&ms, start, stop);

cudaMemcpy(c, dev_c, ROWS*COLUMNS*sizeof(int),cudaMemcpyDeviceToHost);
cudaEventDestroy(start);
cudaEventDestroy(stop);



//cpu_time_used = ((double) (end - start)) / CLOCKS_PER_SEC;
printf("GPU: %f ms",ms);
printf("\n CPU : %ld ms",mtime);
 for (long y = ROWS-1; y < ROWS; y++) // Output Arrays
 {
     for (long x = COLUMNS-1; x < COLUMNS; x++)
     {
      //   printf("\n[%ld][%ld]=%ld ",y,x,c[y][x]);
      //   printf("[%d][%d]=%d ",y,x,d[y][x]);
     }
     printf("\n");
 }
 return 0;
}

输出：

GPU: 0.011040 ms
CPU : 9 ms

现在我可以放心地判断它是否正确吗？

Answer 1

您认为加速太多是正确的，CPU 的时间太长了。使用此方法对 CPU C++ obtaining milliseconds time on Linux -- clock() doesn't seem to work properly 进行计时，您可能还需要将 cudaEventRecord(stop, 0); 移动到内核之后。

我看到你的内核中有 5 次读写。以5*4Bytes*500*500/(1024^3*0.009) 为例，0.517 GB/s 已经从你的记忆中消失了，这只是可用的一小部分。我会说你的 CPU 版本需要一些工作。相比之下，您的 GPU 在5*4Bytes*500*500/(1024^3*0.01104e-3) 大约是421GB/s。我会说你还不在那里。

所以，这么多错误......

#include "stdio.h"
#include <time.h>
#include <sys/time.h>
#include <unistd.h>
#include <cuda.h>
#include <cuda_runtime.h>

#define COLUMNS 500
#define ROWS 500
long a[ROWS*COLUMNS], b[ROWS*COLUMNS], c[ROWS*COLUMNS],d[ROWS*COLUMNS];



__global__ void add(long *a, long *b, long *c,long *d)
{
 int x = blockIdx.x;
 int y = blockIdx.y;
 int i = (COLUMNS*y) + x;
 c[i] = a[i] + b[i];
 a[i]=d[i];
}
int main()
{
 long *dev_a, *dev_b, *dev_c,*dev_d;
 struct timeval startc, end;
 float ms;
 long seconds, useconds;
 double mtime;
 cudaEvent_t start,stop;


 for(int i=0; i<ROWS*COLUMNS; i++)
     d[i]=rand()%4;

 for(int i=0; i<ROWS; i++){
     for(int j=0; j<COLUMNS; j++){
         a[i*COLUMNS+j]=j;
         b[i*COLUMNS+j]=i;
     }
 }

 cudaMalloc((void **) &dev_a, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_b, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_c, ROWS*COLUMNS*sizeof(int));
 cudaMalloc((void **) &dev_d, ROWS*COLUMNS*sizeof(int));



 gettimeofday(&startc, NULL);
 for (long i = 0; i < ROWS*COLUMNS; i++){ // Fill Arrays
     c[i]=a[i]+b[i];
     a[i]=d[i];
 }
  gettimeofday(&end, NULL);

 seconds  = end.tv_sec  - startc.tv_sec;
 useconds = end.tv_usec - startc.tv_usec;
 mtime = useconds;
 mtime/=1000;
 mtime+=seconds*1000;

for (long y = ROWS-1; y < ROWS; y++) // Output Arrays
 {
 for (long x = COLUMNS-1; x < COLUMNS; x++)
 {
    // printf("\n[%ld][%ld]=%ld ",y,x,c[y][x]);
   //   printf("[%d][%d]=%d ",y,x,d[y][x]);
 }
 printf("\n");
 }



cudaMemcpy(dev_a, a, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);
cudaMemcpy(dev_b, b, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);
cudaMemcpy(dev_d, d, ROWS*COLUMNS*sizeof(int),
cudaMemcpyHostToDevice);


cudaEventCreate(&start);
cudaEventCreate(&stop);
cudaEventRecord(start, 0);



add<<<dim3(1024,1024),dim3(128,128)>>>(dev_a, dev_b, dev_c,dev_d);



cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
cudaEventElapsedTime(&ms, start, stop);

cudaMemcpy(c, dev_c, ROWS*COLUMNS*sizeof(int),cudaMemcpyDeviceToHost);
cudaEventDestroy(start);
cudaEventDestroy(stop);

printf("GPUassert: %s\n", cudaGetErrorString(cudaGetLastError()));

//cpu_time_used = ((double) (end - start)) / CLOCKS_PER_SEC;
double memXFers=5*4*COLUMNS*ROWS;
memXFers/=1024*1024*1024;


printf("GPU: %f ms bandwidth %g GB/s",ms, memXFers/(ms/1000.0));
printf("\n CPU : %g ms bandwidth %g GB/s",mtime, memXFers/(mtime/1000.0));
 for (long y = ROWS-1; y < ROWS; y++) // Output Arrays
 {
     for (long x = COLUMNS-1; x < COLUMNS; x++)
     {
      //   printf("\n[%ld][%ld]=%ld ",y,x,c[y][x]);
      //   printf("[%d][%d]=%d ",y,x,d[y][x]);
     }
     printf("\n");
 }

 return 0;
}

顺便说一下我目前的结果（显然不正确）...

GPU: 0.001792 ms bandwidth 2598.56 GB/s
CPU : 0.567 ms bandwidth 8.21272 GB/s