optimization - cudaMallocHost 与 malloc 以获得更好的性能显示没有区别

Question

我经历过这个 site .从这里我得到了使用 cudamallocHost 的固定内存，其性能比 cudamalloc 更好。然后我使用两个不同的简单程序并将执行时间测试为

使用 cudaMallocHost

#include <stdio.h>
#include <cuda.h>

// Kernel that executes on the CUDA device
__global__ void square_array(float *a, int N)
{
  int idx = blockIdx.x * blockDim.x + threadIdx.x;
  if (idx<N) a[idx] = a[idx] * a[idx];
}

// main routine that executes on the host
int main(void)
{
    clock_t start;
    start=clock();/* Line 8 */
    clock_t finish;
  float *a_h, *a_d;  // Pointer to host & device arrays
  const int N = 100000;  // Number of elements in arrays
  size_t size = N * sizeof(float);
  cudaMallocHost((void **) &a_h, size);
  //a_h = (float *)malloc(size);        // Allocate array on host
  cudaMalloc((void **) &a_d, size);   // Allocate array on device
  // Initialize host array and copy it to CUDA device
  for (int i=0; i<N; i++) a_h[i] = (float)i;
  cudaMemcpy(a_d, a_h, size, cudaMemcpyHostToDevice);
  // Do calculation on device:
  int block_size = 4;
  int n_blocks = N/block_size + (N%block_size == 0 ? 0:1);
  square_array <<< n_blocks, block_size >>> (a_d, N);
  // Retrieve result from device and store it in host array
  cudaMemcpy(a_h, a_d, sizeof(float)*N, cudaMemcpyDeviceToHost);
  // Print results
  for (int i=0; i<N; i++) printf("%d %f\n", i, a_h[i]);
  // Cleanup
  cudaFreeHost(a_h);
  cudaFree(a_d);
  finish = clock() - start;
      double interval = finish / (double)CLOCKS_PER_SEC; 
      printf("%f seconds elapsed", interval);
}

使用 malloc

#include <stdio.h>
#include <cuda.h>

// Kernel that executes on the CUDA device
__global__ void square_array(float *a, int N)
{
  int idx = blockIdx.x * blockDim.x + threadIdx.x;
  if (idx<N) a[idx] = a[idx] * a[idx];
}

// main routine that executes on the host
int main(void)
{
    clock_t start;
    start=clock();/* Line 8 */
    clock_t finish;
  float *a_h, *a_d;  // Pointer to host & device arrays
  const int N = 100000;  // Number of elements in arrays
  size_t size = N * sizeof(float);
  a_h = (float *)malloc(size);        // Allocate array on host
  cudaMalloc((void **) &a_d, size);   // Allocate array on device
  // Initialize host array and copy it to CUDA device
  for (int i=0; i<N; i++) a_h[i] = (float)i;
  cudaMemcpy(a_d, a_h, size, cudaMemcpyHostToDevice);
  // Do calculation on device:
  int block_size = 4;
  int n_blocks = N/block_size + (N%block_size == 0 ? 0:1);
  square_array <<< n_blocks, block_size >>> (a_d, N);
  // Retrieve result from device and store it in host array
  cudaMemcpy(a_h, a_d, sizeof(float)*N, cudaMemcpyDeviceToHost);
  // Print results
  for (int i=0; i<N; i++) printf("%d %f\n", i, a_h[i]);
  // Cleanup
  free(a_h); cudaFree(a_d);
  finish = clock() - start;
      double interval = finish / (double)CLOCKS_PER_SEC; 
      printf("%f seconds elapsed", interval);
}

在这两个程序的执行过程中，执行时间几乎相似。
执行上有什么问题吗？？ cudamalloc 和 cudamallochost 在执行上的确切区别是什么？

并且每次运行都会减少执行时间

Answer 1

如果您想查看复制操作的执行时间差异，只需为复制操作计时。在许多情况下，当底层内存被固定时，仅复制操作的执行时间就会有大约 2 倍的差异。并使您的复制操作足够大/足够长，以便您远高于您使用的任何计时机制的粒度。各种profilers例如可视化分析器和 nvprof可以在这里提供帮助。
cudaMallocHost引擎盖下的操作正在做类似 malloc 的事情加上额外的操作系统功能来“固定”与分配相关的每个页面。与仅执行 malloc 相比，这些额外的操作系统操作需要额外的时间。 .请注意，随着分配大小的增加，注册(“固定”)成本通常也会增加。

因此，对于许多示例，仅对整体执行进行计时并不会显示出太大差异，因为虽然 cudaMemcpy从固定内存中操作可能会更快，cudaMallocHost比相应的 malloc 花费的时间更长.

那么有什么意义呢？

当您要从单个缓冲区进行重复传输时，您可能对使用固定内存(即 cudaMallocHost)感兴趣。您只需支付额外的费用即可将其固定一次，但您可以从每次转移/使用中受益。

需要固定内存才能将数据传输操作 ( cudaMemcpyAsync ) 与计算事件(内核调用)重叠。引用 programming guide .