CUDA学习笔记（使用 CPU和GPU 计时器）

在使用CUDA 提升加速计算时，有时会需要和CPU运算的时间做一个对比，以此计算性能提升指标。
在CUDA中，event （事件） API 提供创建和销毁事件、记录事件（包括时间戳）以及将时间戳差异转换为以毫秒为单位的浮点值的调用。

1. CPU计时器
   当计算程序在CPU中执行的时间，可以用#include “time.h”头文件。
   C/C++中的计时函数是clock()，而与其相关的数据类型是clock_t，其中clock_t用来保存时间的数据类型。

clock_t start, finish;	//定义计时开始和结束的时间
start = clock();	//开始计时的时间戳
printf( "helloWord！");;	//执行程序
finish = clock();	//结束计时的时间戳
double duration = (double)(finish - start) / CLOCKS_PER_SEC;	//duration是CPU的计算时间
printf( "%f seconds\n", duration );

注意：常量CLOCKS_PER_SEC，它用来表示一秒钟会有多少个时钟计时单元，其定义如下：#define CLOCKS_PER_SEC ((clock_t)1000)
上面我们看到时钟计时单元的长度为1毫秒，那么计时的精度也为1毫秒，一秒钟有1000个时钟计时单元，而程序执行的单位就是时钟计时单元的个数。所以最后得到的时间值duration的单位应该是秒。

3. GPU计时器
   任何 CPU 计时器均可用于测量 CUDA 调用或内核执行所用的时间。
   使用 CPU 计时器时，务必记住许多 CUDA 函数是异步的；也就是说，它们在完成工作之前将控制权返回给调用 CPU 线程。所有内核启动都是异步的，名字上带有Async后缀的内存复制函数也是如此。因此，要准确测量特定调用或 CUDA 调用序列的耗用时间，必须通过调用将 CPU 线程与 GPU 同步。cudaDeviceSynchronize() 紧接在启动和停止 CPU 定时器之前。 cudaDeviceSynchronize()阻塞调用 CPU 线程，直到该线程先前发出的所有 CUDA 调用完成。

cudaEvent_t start, stop;
float time;

cudaEventCreate(&start);
cudaEventCreate(&stop);

cudaEventRecord( start, 0 );
kernel<<<grid,threads>>> ( d_odata, d_idata, size_x, size_y, 
                           NUM_REPS);
cudaEventRecord( stop, 0 );
cudaEventSynchronize( stop );

cudaEventElapsedTime( &time, start, stop );
cudaEventDestroy( start );
cudaEventDestroy( stop );

这里 cudaEventRecord() 用于放置 start 和 stop事件进入默认流。当它到达流中的事件时，设备将记录事件的时间戳。cudaEventElapsedTime() 函数返回的time参数就是GPU的计算时间。该值以毫秒表示，分辨率约为半微秒。