CUDA 设备到主机复制非常慢答案

【问题标题】：CUDA device to host copy very slowCUDA 设备到主机复制非常慢
【发布时间】：2012-09-29 08:25:42
【问题描述】：

我正在运行 Windows 7 64 位、cuda 4.2、Visual Studio 2010。

首先，我在 cuda 上运行一些代码，然后将数据下载回主机。然后进行一些处理并移回设备。然后我做了以下从设备到主机的复制，它运行得非常快，比如 1 毫秒。

clock_t start, end;
count=1000000;
thrust::host_vector <int> h_a(count);
thrust::device_vector <int> d_b(count,0);
int *d_bPtr = thrust::raw_pointer_cast(&d_b[0]);
start=clock();
thrust::copy(d_b.begin(), d_b.end(), h_a.begin());
end=clock();
cout<<"Time Spent:"<<end-start<<endl;

大约需要 1 毫秒才能完成。

然后我再次在 cuda 上运行了一些其他代码，主要是原子操作。然后我将数据从设备复制到主机，这需要很长时间，大约9s。

__global__ void dosomething(int *d_bPtr)
{
....
atomicExch(d_bPtr,c)
....
}

start=clock();
thrust::copy(d_b.begin(), d_b.end(), h_a.begin());
end=clock();
cout<<"Time Spent:"<<end-start<<endl;

~9s

我多次运行代码，例如

int i=0;
while (i<10)
{
clock_t start, end;
count=1000000;
thrust::host_vector <int> h_a(count);
thrust::device_vector <int> d_b(count,0);
int *d_bPtr = thrust::raw_pointer_cast(&d_b[0]);
start=clock();
thrust::copy(d_b.begin(), d_b.end(), h_a.begin());
end=clock();
cout<<"Time Spent:"<<end-start<<endl;

__global__ void dosomething(int *d_bPtr)
{
....
atomicExch(d_bPtr,c)
....
}

start=clock();
thrust::copy(d_b.begin(), d_b.end(), h_a.begin());
end=clock();
cout<<"Time Spent:"<<end-start<<endl;
i++
}

结果几乎相同。
可能是什么问题？

谢谢！

【问题讨论】：

我很好奇，你是在主机端编程还是设备函数调用 atomicExch？ atomicExch 只能在设备函数/内核中调用。
我仍然不明白你如何使用thrust::raw_ptr_cast 和device_vector 第一个索引。我正在尝试从你的代码运行一个sn-p，我收到error: argument list for class template "thrust::device_ptr" is missing 错误.. .
对不起，我的错。它应该是 int *device_ptr =thrust::raw_pointer_cast(&d_b[0]);我会更新它。你认为这是造成问题的原因吗？还是应该直接使用 d_b.begin() 作为原子操作的输入？谢谢！
你能发布你能想到的最短的复制器吗？我尝试从您的代码中制作一个简单的示例，但没有发现任何问题。您的代码中存在各种奇怪的语法错误，因此拥有一个可编译的复制器会有所帮助。
真的很抱歉，这是我的错。我没有带我的源代码。我会按照 talonmies 的建议，重新运行测试。并且明天尽快发布一个可编译的代码。非常感谢！！

标签： c++ performance cuda copy device

【解决方案1】：

问题是时间问题，而不是复制性能的任何变化。内核启动在 CUDA 中是异步的，因此您测量的不仅仅是thrust::copy 的时间，还包括您启动的先前内核完成的时间。如果您将用于计时复制操作的代码更改为如下所示：

cudaDeviceSynchronize(); // wait until prior kernel is finished
start=clock();
thrust::copy(d_b.begin(), d_b.end(), h_a.begin());
end=clock();
cout<<"Time Spent:"<<end-start<<endl;

您应该会发现传输时间已恢复到之前的性能。所以你真正的问题不是“为什么thrust::copy 慢”，而是“为什么我的内核慢”。根据您发布的相当糟糕的伪代码，答案是“因为它充满了序列化内核内存事务的atomicExch() 调用”。

【讨论】：

谢谢你的爪子！！我会按照你的建议，明天再运行代码看看。对不起可怕的伪代码。我是 cuda 新手，现在没有可用的源代码...非常感谢！
我今天测试了，talonmies 完全正确！！一切都和他描述的一样！！非常感谢！
@UserKiwi：如果这回答了你的问题，那么也许你会对accept it 这么好。这标志着问题已完成。

【解决方案2】：

我建议你使用 cudpp，在我看来比推力更快（我正在写关于优化的硕士论文，我尝试了这两个库）。如果拷贝很慢，可以尝试自己写内核拷贝数据。

【讨论】：