我的电脑在处理大向量时，CUDA Thrust操作变得缓慢

Question

我是一个CUDA初学者，在阅读一些thrust教程。我写了一段简单但十分混乱的代码，并尝试找出如何加速thrust。（这个想法正确吗？）我试图将两个向量（每个向量有10000000个int）添加到另一个向量中，使用cpu上的数组相加和gpu上的device_vector相加。

以下是要注意的事项：

#include <iostream>
#include "cuda.h"
#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <thrust/device_vector.h>
#include <thrust/host_vector.h>

#define N 10000000
int main(void)
{
    float time_cpu;
    float time_gpu;
    int *a = new int[N];
    int *b = new int[N];
    int *c = new int[N];
    for(int i=0;i<N;i++)
    {
        a[i]=i;
        b[i]=i*i;
    }
    clock_t start_cpu,stop_cpu;
    start_cpu=clock();
    for(int i=0;i<N;i++)
    {
        c[i]=a[i]+b[i];
    }
    stop_cpu=clock();   
    time_cpu=(double)(stop_cpu-start_cpu)/CLOCKS_PER_SEC*1000;
    std::cout<<"Time to generate (CPU):"<<time_cpu<<std::endl;
    thrust::device_vector<int> X(N);
    thrust::device_vector<int> Y(N);
    thrust::device_vector<int> Z(N);
    for(int i=0;i<N;i++)
    {
        X[i]=i;
        Y[i]=i*i;
    }
    cudaEvent_t start, stop;
    cudaEventCreate(&start);
    cudaEventCreate(&stop);
    cudaEventRecord(start,0);       
    thrust::transform(X.begin(), X.end(),
        Y.begin(),
        Z.begin(),
        thrust::plus<int>());
    cudaEventRecord(stop,0);
    cudaEventSynchronize(stop);
    float elapsedTime;
    cudaEventElapsedTime(&elapsedTime,start,stop);
    std::cout<<"Time to generate (thrust):"<<elapsedTime<<std::endl;
    cudaEventDestroy(start);
    cudaEventDestroy(stop); 
    getchar();
    return 0;
}

CPU的结果非常快，但是在我的机器上GPU运行非常慢（i5-2320，4G，GTX 560 Ti），CPU时间约为26，GPU时间约为30！我是不是在代码中做了愚蠢的thrust错误或者有更深层次的原因？

作为一个C++新手，我反复检查我的代码，使用thrust仍然得到了较慢的GPU时间，因此我进行了一些实验，展示了使用五种不同方法计算vectorAdd的差异。我使用Windows API QueryPerformanceFrequency() 作为统一的时间测量方法。

每个实验看起来都像这样：

f = large_interger.QuadPart;  
QueryPerformanceCounter(&large_interger);  
c1 = large_interger.QuadPart; 

for(int j=0;j<10;j++)
{
    for(int i=0;i<N;i++)//CPU array adding
    {
        c[i]=a[i]+b[i];
    }
}
QueryPerformanceCounter(&large_interger);  
c2 = large_interger.QuadPart;  
printf("Time to generate (CPU array adding) %lf ms\n", (c2 - c1) * 1000 / f);

这是我的一个简单的__global__函数，用于GPU数组相加：

__global__ void add(int *a, int *b, int *c)
{
    int tid=threadIdx.x+blockIdx.x*blockDim.x;
    while(tid<N)
    {
        c[tid]=a[tid]+b[tid];
        tid+=blockDim.x*gridDim.x;
    }
}

函数被称为：

for(int j=0;j<10;j++)
{
    add<<<(N+127)/128,128>>>(dev_a,dev_b,dev_c);//GPU array adding
}   

我使用以下循环十次将向量a[N]和b[N]加入到向量c[N]中：

1. 在CPU上添加数组
2. 在CPU上添加std::vector
3. 在CPU上添加thrust::host_vector
4. 在GPU上添加thrust::device_vector
5. 在GPU上添加数组。这是结果

其中N=10000000。

我得到的结果如下:

1. CPU数组添加时间：268.992968毫秒
2. CPU std::vector添加时间：1908.013595毫秒
3. CPU Thrust::host_vector添加时间: 10776.456803毫秒
4. GPU Thrust::device_vector添加时间: 297.156610毫秒
5. GPU数组添加时间: 5.210573毫秒

这让我感到困惑，因为我不熟悉模板库的实现。容器和原始数据结构之间的性能真的会有这么大的差异吗？

- Tony

Answer 1

大部分执行时间都花费在初始化X[i]和Y[i]的循环中。虽然这样做是合法的，但这种方式初始化大型设备向量非常慢。最好创建主机向量、对其进行初始化，然后将其复制到设备上。作为一个测试，在你初始化设备向量X[i]和Y[i]的循环之后，修改你的代码如下：

}  // this is your line of code
std::cout<< "Starting GPU run" <<std::endl;  //add this line
cudaEvent_t start, stop;   //this is your line of code

您会看到，添加该行代码后，GPU计时结果几乎立即出现。 因此，您等待的所有时间都花费在直接从主机代码初始化这些设备向量上。

当我在我的笔记本电脑上运行此代码时，CPU时间约为40，GPU时间约为5，因此GPU在实际进行计时的代码段中运行速度约为CPU的8倍。

如果您将X和Y创建为主机向量，然后创建类似的d_X和d_Y设备向量，则整体执行时间将缩短，如下所示：

thrust::host_vector<int> X(N);     
thrust::host_vector<int> Y(N);     
thrust::device_vector<int> Z(N);     
for(int i=0;i<N;i++)     
{     
    X[i]=i;     
    Y[i]=i*i;     
}   
thrust::device_vector<int> d_X = X;
thrust::device_vector<int> d_Y = Y;

并将您的转换调用更改为：

thrust::transform(d_X.begin(), d_X.end(),      
    d_Y.begin(),      
    Z.begin(),      
    thrust::plus<int>()); 

好的，现在你已经表明 CPU 运行速度比 GPU 测量速度更快了。很抱歉我先前误解了。我的笔记本电脑是一台带有2.6GHz core i7和Quadro 1000M gpu的惠普笔记本电脑。我正在运行centos 6.2 linux。几点评论：如果您在GPU上运行任何重负载的显示任务，那么这可能会影响性能。此外，在对这些内容进行基准测试时，通常使用相同的比较机制。您可以同时使用cudaEvents来计时CPU代码和GPU代码。此外，通常使用thrust进行一次未计时的热身测试，然后重复测试以进行测量。同样，通常在循环中运行测试10次或更多次，然后除以获取平均值。在我的情况下，我可以发现clocks（）测量结果相当粗糙，因为连续的运行会给出30、40或50。在GPU测量中，我得到了大约5.18256的值。其中一些东西可能会有所帮助，但我无法确切地说您的结果与我的结果在GPU方面如此不同的原因是什么。

好的，我又做了一个实验。编译器对CPU方面有很大影响。我使用 -O3 开关进行编译，CPU时间降至0。然后我将CPU计时测量从 clocks（）方法转换为 cudaEvents，得到CPU测量时间为12.4（使用 -O3 优化），GPU方面仍为5.1。

根据计时方法和在CPU方面使用的编译器，您的结果可能有所不同。

- Robert Crovella

---

我没有看到他计时初始化部分，所以我不认为那是问题。 - Pavan Yalamanchili

当你实际运行代码时，时间显示出合理的数字，即报告的GPU时间比报告的CPU时间快，正如我在回答中提到的那样。我认为这也不是问题所在。我相信OP感到困惑是因为整个执行时间很长。 - Robert Crovella

我知道初始化部分可能会非常慢，感谢您创建host_vector的建议。但问题是在我的电脑上，CPU时间约为26，GPU时间约为30！（对不起，我在我的问题中没有讲清楚，我已经编辑了）我还改变了Y [i] = i * i和c [i] = i * i为Y [i] = i和c [i] = i。奇怪的是，我在想GPU时间是否被乘以10倍...您是如何在笔记本电脑上运行代码的？@ Robert @gpu - Tony

在我的“答案”帖子中添加了一些回复。 - Robert Crovella

@Robert 我也遇到了0毫秒的问题。我也做了一些实验并编辑了我的问题。你可以看一下。 - Tony

Answer 2

首先，Y[i]=i*i; 在10M个元素中不适用于整数。整数大约可以容纳1e10，而您的代码需要1e14。

其次，看起来变换的时间是正确的，并且应该比CPU更快，无论您使用哪个库。对于这种情况，Robert建议在CPU上初始化向量，然后将其传输到GPU，这是一个好主意。

第三，由于我们无法进行整数倍增，下面是一些更简单的CUDA库代码（使用我正在开发的ArrayFire）以类似的方式处理浮点数，供您进行基准测试：

int n = 10e6;
array x = array(seq(n));
array y = x * x;
timer t = timer::tic();
array z = x + y;
af::eval(z); af::sync();
printf("elapsed seconds: %g\n", timer::toc( t));

祝你好运！

- arrayfire

Answer 3

最近我正在使用CUDA Thrust在我的Quadro 1000m上运行类似的测试。我使用thrust::sort_by_key作为基准来测试其性能，结果太好了，让我的老板信服了。对于512MB的配对，它需要100+毫秒进行排序。

对于您的问题，我有两个困惑。

（1）为什么要将time_cpu乘以1000？没有1000，它已经是秒了。

time_cpu=(double)(stop_cpu-start_cpu)/CLOCKS_PER_SEC*1000;

（2）并且，当您提到26、30、40时，您是指秒还是毫秒？ 'cudaEvent'报告经过的时间是以“ms”而不是“s”为单位。

- thierry