小编Tre*_*man的帖子

auto ts = std::chrono::system_clock::now();

cudaMemcpyAsync((void**)in_dev, in_host, 1000 * size, cudaMemcpyHostToDevice, stream_in);
cudaMemcpyAsync((void**)out_host, out_dev, 1000 * size, cudaMemcpyDeviceToHost, stream_out);

cudaStreamSynchronize(stream_in);
cudaStreamSynchronize(stream_out);

time_data.push_back(std::chrono::system_clock::now() - ts);

这是我为自己的教育目的制定的基准的结果。非常简单，程序的每个“周期”都会启动并行数据传输，并在获取时间戳之前等待这些操作完成。

内核版本添加了一个简单的内核，该内核对每个数据字节（也在不同的流上）进行操作。内核执行时间的趋势对我来说很有意义 - 我的设备只有这么多 SM/核心，一旦我要求更多，它就会开始花费更长的时间。

我不明白的是，为什么仅内存传输测试在与核心限制几乎相同的数据大小点上开始呈指数级增长。我的设备的内存带宽标榜为 600 GB/s。此处传输 10 MB 平均需要约 1.5 毫秒，这并不是给定带宽的餐巾纸数学建议的结果。我的预期是内存传输延迟周围的时间几乎是恒定的，但情况似乎并非如此。

为了确认这不是我的盗版时间戳方法，我使用 NSight Compute 运行了仅内存版本，并确认从 N=1000 KB 到 N=10000 KB 将平均异步传输时间从约 80 us 增加到约 800 us。

我对 D/H 内存传输性能缺少什么？获得良好带宽的关键是重叠大量小型传输而不是大型传输，还是会因为有限的复制引擎瓶颈而变得更糟？

我在配备 PCIe4 系统的 RTX 3070 Ti 上运行了此基准测试。