为什么 cudaMemcpy 花费这么多时间？答案

【问题标题】：Why cudaMemcpy cost so much time?为什么 cudaMemcpy 花费这么多时间？
【发布时间】：2014-06-08 18:20:19
【问题描述】：

我正在编写 cuda 程序，并且在分析了一项功能（例如在大型矩阵上进行点积）后，大部分时间都需要花费：

==27530== API calls:
Time(%)      Time     Calls       Avg       Min       Max  Name
 64.90%  2.25369s        23  97.986ms  9.5590us  1.79533s  cudaMemcpy
 21.04%  730.65ms      1422  513.82us  3.0050us  21.028ms  cudaLaunch
  8.72%  302.72ms         5  60.543ms     477ns  170.92ms  cudaFree
  3.64%  126.54ms        18  7.0298ms  4.8882ms  35.518ms  cudaMallocHost
  1.39%  48.292ms        16  3.0182ms  3.0076ms  3.0601ms  cudaFreeHost
  0.11%  3.9026ms        23  169.68us  64.314us  1.7771ms  cudaMalloc
  0.09%  3.0171ms     17661     170ns     144ns  3.1750us  cudaSetupArgument
  0.04%  1.3514ms       810  1.6680us  1.4000us  9.9270us  cudaBindTexture
  0.02%  569.60us       810     703ns     596ns  4.8010us  cudaUnbindTexture
  0.02%  556.24us       945     588ns     484ns  4.2560us  cudaFuncSetCacheConfig
  0.01%  499.67us      1422     351ns     163ns  198.52us  cudaConfigureCall
  0.01%  256.21us      1310     195ns     150ns     335ns  cudaGetLastError
  0.01%  238.26us       166  1.4350us     165ns  49.141us  cuDeviceGetAttribute
  0.01%  175.44us       945     185ns     157ns     755ns  cudaPeekAtLastError
  0.00%  50.787us         2  25.393us  16.700us  34.087us  cuDeviceGetName
  0.00%  45.330us         2  22.665us  19.024us  26.306us  cuDeviceTotalMem
  0.00%  43.289us         2  21.644us  13.641us  29.648us  cudaMemset
  0.00%  43.029us         2  21.514us  14.059us  28.970us  cudaGetDeviceProperties
  0.00%  13.931us        12  1.1600us     339ns  5.5310us  cudaGetDevice
  0.00%  3.4750us         1  3.4750us  3.4750us  3.4750us  cudaDeviceSynchronize
  0.00%  1.5320us         1  1.5320us  1.5320us  1.5320us  cuDriverGetVersion
  0.00%  1.2690us         3     423ns     241ns     753ns  cuDeviceGetCount
  0.00%  1.0080us         1  1.0080us  1.0080us  1.0080us  cuInit
  0.00%  1.0060us         3     335ns     314ns     377ns  cuDeviceGet

它显示“cudaMemcpy”的花费大约超过两秒。但是我的代码中很少有 cudaMemcpy 调用，并且 D->H 或 H->D 内存副本都是固定内存。我认为我的 cudaMemcpy 调用不会花费这么多时间。

消耗时间最多的函数：

==27530== Profiling result:
Time(%)      Time     Calls       Avg       Min       Max  Name
 74.35%  2.34598s       112  20.946ms  20.743ms  21.161ms  knl_convolve_filter(float*, float*, int, int, int, float*)

和功能：

__global__ void knl_convolve_filter(float *feature, float *filter, int width, int height, int cell_size, float *convolution) {
    int x =  blockDim.x * blockIdx.x + threadIdx.x;
    int y =  blockDim.y * blockIdx.y + threadIdx.y;

    if( x < width && y < height) {
        if( x & 1) {
            //odd, imaginary part
            float sum = 0.0f;
            size_t offset = (y * width + x - 1) * cell_size ;
            for(int i = 0, total_cell_size = cell_size * 2; i < total_cell_size ; i += 2) {
                float y = *(feature + offset + i) * *(filter + offset + i + 1) + *(feature + offset + i + 1) * *(filter + offset + i);
                sum += y;
            }
            *(convolution + y * width + x) = sum;
        } else {
            //even, real part
            float sum = 0.0f;
            size_t offset = (y * width + x) * cell_size ;
            for(int i = 0, total_cell_size = cell_size * 2; i < total_cell_size ; i += 2) {
                float x = *(feature + offset + i) * *(filter + offset + i) - *(feature + offset + i + 1) * *(filter + offset + i + 1);
                sum += x;
            }
            *(convolution + y * width + x) = sum;
        }

    }
}

我在 Fedora 19 64、cuda 6.0 上使用 GTX760(CC3.0)。我在这里做错了什么吗？

【问题讨论】：

我不熟悉 cuda，但您的 knl_convolve_filter 函数看起来不像调用 cuda 函数？另外，请注意，一个 cudaMemcpy 调用需要 1.79 秒，所以这可能是扭曲的东西？
重点是 one cudaMemcpy 调用耗时 1.79533s。这几乎可以肯定意味着它是程序中的第一个运行时 API 调用，并且会导致运行时 API 的所有延迟初始化和设置成本的损失。
删除内核调用可能会减少运行时初始化的延迟。如果没有一个简短、完整的例子来说明问题，就不可能告诉你观察到的“缓慢”的确切来源是什么
在 main 函数的开头粘贴 cudaFree(0);。然后重新配置文件。您可能会得到更合理的结果，cudaFree 操作会因运行时初始化而吸收大部分延迟。
@RobertCrovella：不要停止提问，但也许请记住为什么我在 cmets 中为很多问题提供“几乎”答案。请注意，无论如何我确实回答了这个问题，因此投票会将其从未回答列表中移出....

标签： c++ cuda

【解决方案1】：

很难给出明确的答案，因为我们没有看到任何主机代码，但事实上，一个在分析中似乎有一个非常慢的 cudaMemcpy 调用耗时 1.79533 秒的序列。其他 20 次调用平均只需要大约 20 毫秒。所以真正的问题是“为什么这个特定的cudaMemcpy 调用需要 1.79533 秒？”，我怀疑答案是它吸收了 CUDA 运行时 API 中的大量延迟设置延迟。

现代版本的 CUDA 工具包附带的 nvprof 配置文件实用程序可以选择发出详细的 API 时间线。对该时间线的分析肯定会回答您的问题，但在没有主机代码或 API 跟踪的情况下，这是可以提供的尽可能具体的答案。

【讨论】：