C++、CUDA程序计时

最近在测试自己写的cuda程序性能，需要用到计时，简单总结了以下查阅到的计时方法。如有补充，欢迎评论、私信交流。

计时一般分两种情况：

（1）测c++、CUDA混合编码（或者说同时包含设备代码、主机代码的混合代码）的程序段；

（2）只测核函数、设备内存拷贝函数等设备代码。

对于第（1）种情况，常用的计时方式有两种（欢迎补充），都是库函数里提供的方法。其次还有一种方法也能使用，不过需要注意使用场合。

对于第（2）种情况，一般用nsys指令等。

①gettimeofday()

它的精度可以达到微秒。

注意：核函数的执行相对cpu来说是异步执行的，所以必须加上核函数同步函数，才能得到准确的时间。

#include<stdio.h>
#include <stdlib.h>
#include<sys/time.h>

double cpuSecond() {
    struct timeval tp;
    gettimeofday(&tp,NULL);//在使用gettimeofday()函数时，第二个参数一般都为空，因为我们一般都只是为了获得当前时间，而不用获得timezone的数值
    return ((double)tp.tv_sec + (double)tp.tv_usec*1.e-6);
}

int main()
{
    double iStart = cpuSecond();//开始时间
    
    /*一些运行代码*/

    cudaDeviceSynchronize();  // 同步函数
    double iElaps = cpuSecond() - iStart;//结束时间-开始时间
    return 0;
}

②clock()

计时的精度为毫秒。

#include <stdio.h>
#include <stdlib.h>
#include <time.h>//实际上我在使用该方法时，未include该库，也能使用    

int main()
{
    clock_t start,stop;
    float time;
    start clock();//开始时间

    /*需要计时的代码*/

    cudaDeviceSynchronize();  // 同步函数
    stop clock();//结束时间
    time = (stop - start)/CLOCK_PER_SEC;
    return 0;
}

③计时事件函数

计算两次事件之间相差的时间（以毫秒为单位，精度为0.5微秒）。

CUDA中的事件本质上是一个GPU时间戳，这个时间戳是在用户指定的时间点上记录的。由于GPU本身支持记录时间戳，因此就避免了当使用CPU定时器来统计GPU执行的时间时可能遇到的诸多问题。

注意：开始计时之后紧接着GPU执行的函数，停止计时前紧接着GPU执行的函数（中间的步骤可以有主机代码），否则记录会不准确。

//无需include额外的库

int main()
{
    cudaEvent_t start, stop;
	float elapsedTime = 0.0;

	cudaEventCreate(&start);
	cudaEventCreate(&stop);
	cudaEventRecord(start, 0);//开始时间

    /*运行的代码*/

    cudaEventRecord(stop, 0);
	cudaEventSynchronize(stop);//结束时间
	cudaEventElapsedTime(&elapsedTime, start, stop);

	cout << elapsedTime <<"ms"<< endl;

	cudaEventDestroy(start);
	cudaEventDestroy(stop);
    return 0；
} 

如何获得精确的计时

正常情况下，第一次执行核函数的时间会比第二次慢一些。这是因为GPU在第一次计算时需要warmup。若使用了Opencv的CUDA接口，实践中发现也需要warmup。GPU的warmup就是指在计时前，先在设备端跑一个空的核函数，预热一下主机与设备端之间的数据连接等。

__global__ void warmup(){}

int main()
{
warmup<<<1,1>>>();
//计时开始
...
//计时结束
return 0;
}

参考博客链接：CUDA进阶第三篇：CUDA计时方式_cuda warmup_大光叔叔的博客-CSDN博客