创建CUDA上下文的区别

Question

我有一个使用三个内核的程序。为了获得加速，我进行了虚拟内存复制以创建上下文，具体操作如下：

__global__ void warmStart(int* f)
{
    *f = 0;
}

我想对以下在内核之前启动的程序进行计时：

int *dFlag = NULL;
cudaMalloc( (void**)&dFlag, sizeof(int) );
warmStart<<<1, 1>>>(dFlag);
Check_CUDA_Error("warmStart kernel");

我还了解到其他最简单的创建上下文方式，如cudaFree(0)或cudaDeviceSynchronize()。但是使用这些API调用会导致比使用虚拟内核更差的执行时间。

在强制创建上下文后，程序的执行时间为0.000031秒（虚拟内核）和0.000064秒（cudaDeviceSynchronize()和cudaFree(0)都是），这些时间是对程序进行10次单独执行的平均值。

因此，我得出的结论是，在规范的方式下创建上下文时，启动一个内核会初始化某些未初始化的内容。

那么，使用内核和使用API调用创建上下文有什么区别呢？

我在Linux下使用CUDA 4.0在GTX480上运行了这个测试。

- pQB

Answer 1

每个CUDA上下文都有必须执行内核所需的内存分配，这些内存分配不需要用于同步、分配内存或释放内存。上下文内存的初始分配和这些分配的重新调整被推迟到内核需要这些资源时。这些分配的示例包括本地内存缓冲区、设备堆栈和printf堆。

- Greg Smith