用 cudaMemPrefetchAsync 替换 cudaMemcpy2D

Question

我正在尝试以相同的方式对驻留在统一内存中的数据进行异步内存传输（主机到设备） cudaMemcpy2DAsync() 提供，即使用 音高值 作为目标（设备）内存。

但是，据我了解：

我无法使用 cudaMemcpy2DAsync()，因为它需要固定内存，而使用 cudaMallocManaged() 分配的统一内存块默认情况下未固定，无法手动固定。
我不能使用 cudaMemPrefetchAsync() 因为它只复制连续的内存块，而 cudaMemcpy2DAsync() 包括额外的选项来根据给定的间距值缓冲我的数据行.

我的问题：

如何传输我的数据，(a) 驻留在统一内存，(b) 从 Host 到 Device 的异步方式，同时确保数据被复制 (c) 具有正确的音高值?

我目前使用 cudaMemcpy2D() 进行的传输看起来像这样：

cudaMemcpy2D( 
    dest_ptr, dest_pitch,         // dst address & pitch
    src_ptr, dim_x*sizeof(float)  // src address & pitch
    dim_x*sizeof(float), dim_y,   // transfer width & height
    cudaMemcpyHostToDevice ) );

（如您所见，源头的音高实际上为零，而目的地的音高为 dest_pitch -- 也许这有帮助？）

一个额外的麻烦是我不分配需要自己传输的数据，所以我不能在不创建额外的数据副本的情况下手动应用音调（这会有问题）。

我的一个想法 是简单地逐行复制数据。然而，这将意味着大量非常小的数据传输，这听起来有点可怕。在我的示例中，它将是 3,040,000 次传输，每次传输 304 个字节。但如果传输异步分布在多个流中，那么这在 Pascal 上可能没问题..?

任何 pointers/ideas 将不胜感激！

Answer 1

答案是您需要复印一份。统一内存的异步副本不支持pitches，因为统一内存不支持pitches。您可以将设备上预取的统一缓冲区复制到一个倾斜的内存区域，也可以将主机上的统一缓冲区复制到一个倾斜的内存区域，然后异步复制它。您不能同时满足所有三个条件（a、b、c）

用 cudaMemPrefetchAsync 替换 cudaMemcpy2D

Replacing cudaMemcpy2D with cudaMemPrefetchAsync

c++

cuda

unified-memory