CUDA学习--页锁定主机内存

1. 页锁定主机内存

除了通过cudaMalloc()在GPU上分配内存，以及通过标准的C函数malloc()在主机上分配内存，CUDA运行时还提供了自己独有的机制来分配主机内存：cudaHostAlloc()。

C函数malloc()将分配标准的，可分页的主机内存。而cudaHostAlloc()将分配页锁定的主机内存。页锁定的主机内存也称为固定内存或不可分页内存，它的重要属性就是：操作系统将不会对这块内存分页并交换到磁盘上，从而确保了该内存始终驻留在物理内存中。因此，操作系统能够安全的使用应用程序访问该内存的物理地址，因为这块内存将不会被破坏或者重新定位。
事实上，当使用可分页内存进行复制时，复制操作将执行两遍，第一遍从可分页内存复制到一块“临时的”页锁定内存，然后再从这个页锁定内存复制到GPU上。因此，当在GPU和主机间复制数据时，这种差异会使也锁定主机内存的性能比标准可分页内存的性能要高大约2倍。
然而，我们也不能进入另一个极端：查找每一个malloc调用并将其替换为cudaHostAlloc调用。固定内存是一把双刃剑，当使用固定内存是，你将失去虚拟内存的所用功能。特别是，应用程序中使用每个页锁定内存时都需要分配物理内存，因为这些内存不能交换到磁盘上。这意味着，与使用标准的malloc调用相比，系统将更快的耗尽内存。因此，建议仅对cudaMemcpy()调用中的源内存或者目标内存，才使用页锁定内存，并且在不需要的时候立即释放。
页锁定内存的分配与释放：

float cuda_host_alloc_test(int size, bool up){cudaEvent_t start, stop;int *a, *dev_a;float elapsedTime;cudaEventCreate(&start);cudaEventCreate(&stop);cudaError_t cudaStatus = cudaHostAlloc((void **)&a, size * sizeof(*a), cudaHostAllocDefault);if (cudaStatus != cudaSuccess){printf("host alloc fail!\n");return -1;}cudaStatus = cudaMalloc((void **)&dev_a, size * sizeof(*dev_a));if (cudaStatus != cudaSuccess){fprintf(stderr, "cudaMalloc failed!\n");return -1;}cudaEventRecord(start, 0);for (int i = 0; i < 100; i++){if (up){cudaStatus = cudaMemcpy(dev_a, a, size * sizeof(*dev_a), cudaMemcpyHostToDevice);if (cudaStatus != cudaSuccess){fprintf(stderr, "cudaMemcpy Host to Device failed!\n");return -1;}}else{cudaStatus = cudaMemcpy(a, dev_a, size * sizeof(*dev_a), cudaMemcpyDeviceToHost);if (cudaStatus != cudaSuccess){fprintf(stderr, "cudaMemcpy Device to Host failed!\n");return -1;}}}cudaEventRecord(stop, 0);cudaEventSynchronize(stop);cudaEventElapsedTime(&elapsedTime, start, stop);cudaFreeHost(a);cudaFree(dev_a);cudaEventDestroy(start);cudaEventDestroy(stop);return elapsedTime;}

可以看到，cudaHostAlloc分配的内存与malloc()分配的内存在使用方式上是相同的。与malloc()的不同之处在于

最后一个参数cudaHostAllocDefault。最后一个参数的取值范围是一组枚举量，我们可以通过这些变量来修改

cudaHostAlloc的行为。

函数由参数up来制定方向，对内存复制进行了一百次操作后，释放主机缓冲区和GPU缓冲区，并销毁计时器。