CUDA 设备到主机的复制非常慢

Question

CUDA 设备到主机的复制非常慢

4

我正在使用Windows 7 64位操作系统，CUDA 4.2和Visual Studio 2010。

首先，我在CUDA上运行了一些代码，然后将数据下载回主机。接着进行一些处理并移回设备。然后我从设备向主机复制，速度非常快，大约只需要1ms。

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上运行了一些其他的代码，主要是原子操作。然后我将数据从设备复制到主机，这需要很长时间，大约需要9秒。

__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;

~ 9秒

我多次运行了代码，例如：

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++
}

结果基本相同。可能的问题是什么？谢谢！

- UserKiwi

我很好奇，你是在主机端编程还是设备函数中调用atomicExch？atomicExch只能在设备函数/内核中调用。 - Hong Zhou

我仍然不明白你是如何能够在device_vector的第一个索引中使用thrust::raw_ptr_cast。我正在尝试运行你代码中的片段，但是我遇到了“error: argument list for class template "thrust::device_ptr" is missing”错误。 - Recker

对不起，我的错。应该是int *device_ptr = thrust::raw_pointer_cast(&d_b[0]); 我会更新的。你认为是这个引起了问题吗？还是我直接使用d_b.begin()作为原子操作的输入？谢谢！ - UserKiwi

你能提供一个最短的可重现代码吗？我试着从你的代码中制作了一个简单的例子，但没有发现任何问题。你的代码中有各种奇怪的语法错误，因此提供一个可编译的可重现代码将会很有帮助。 - Robert Crovella

非常抱歉，这是我的错。我没有带我的源代码。我会遵循talonmies的建议，重新运行测试。并且明天尽快发布可编译的代码。非常感谢！ - UserKiwi

2个回答

-2

我建议你使用cudpp，在我看来比thrust快（我正在写有关优化的硕士论文，并尝试了这两个库）。如果复制非常慢，你可以尝试编写自己的内核来复制数据。

- Tomasz Dzięcielewski

网页内容由stack overflow 提供, 点击上面的

可以查看英文原文，
原文链接

- talonmies · Accepted Answer

问题在于时间上的因素，而不是拷贝性能的任何变化。在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()调用，这些调用会序列化内核中的内存事务”。