28

我正在寻找可以为 CPU（使用 g++）和 GPU（使用 nvcc）编写的最简洁的代码量，GPU 的性能始终优于 CPU。任何类型的算法都是可以接受的。

澄清一下：我实际上是在寻找两个短代码块，一个用于 CPU（在 g++ 中使用 C++），另一个用于 GPU（在 nvcc 中使用 C++），GPU 的性能优于该代码块。最好以秒或毫秒为单位。可能的最短代码对。

4

4 回答 4

41

首先，我将重申我的评论：GPU 是高带宽、高延迟的。试图让 GPU 在纳秒的工作（甚至是毫秒或秒的工作）上击败 CPU 完全失去了做 GPU 工作的意义。下面是一些简单的代码，但要真正体会到 GPU 的性能优势，你需要一个很大的问题规模来分摊启动成本……否则，它就毫无意义。我可以在两英尺的比赛中击败法拉利，仅仅是因为转动钥匙、启动发动机和踩踏板需要一些时间。这并不意味着我在任何有意义的方面都比法拉利更快。

在 C++ 中使用类似的东西：

  #define N (1024*1024)
  #define M (1000000)
  int main()
  {
     float data[N]; int count = 0;
     for(int i = 0; i < N; i++)
     {
        data[i] = 1.0f * i / N;
        for(int j = 0; j < M; j++)
        {
           data[i] = data[i] * data[i] - 0.25f;
        }
     }
     int sel;
     printf("Enter an index: ");
     scanf("%d", &sel);
     printf("data[%d] = %f\n", sel, data[sel]);
  }

在 CUDA/C 中使用类似的东西：

  #define N (1024*1024)
  #define M (1000000)

  __global__ void cudakernel(float *buf)
  {
     int i = threadIdx.x + blockIdx.x * blockDim.x;
     buf[i] = 1.0f * i / N;
     for(int j = 0; j < M; j++)
        buf[i] = buf[i] * buf[i] - 0.25f;
  }

  int main()
  {
     float data[N]; int count = 0;
     float *d_data;
     cudaMalloc(&d_data, N * sizeof(float));
     cudakernel<<<N/256, 256>>>(d_data);
     cudaMemcpy(data, d_data, N * sizeof(float), cudaMemcpyDeviceToHost);
     cudaFree(d_data); 

     int sel;
     printf("Enter an index: ");
     scanf("%d", &sel);
     printf("data[%d] = %f\n", sel, data[sel]);
  }

如果这不起作用，请尝试使 N 和 M 更大，或将 256 更改为 128 或 512。

于 2011-10-05T16:03:34.723 回答
4

一个非常非常简单的方法是计算前 100,000 个整数的平方，或者一个大型矩阵运算。Ita 易于实现，并且通过避免分支、不需要堆栈等来利用 GPU 的优势。我不久前使用 OpenCL 与 C++ 进行了此操作，并获得了一些非常惊人的结果。（2GB GTX460 的性能大约是双核 CPU 的 40 倍。）

您是在寻找示例代码，还是只是想法？

编辑

40x 与双核 CPU 相比，而不是四核。

一些指示：

  • 确保您在运行基准测试时没有运行，比如说，孤岛危机。
  • 关闭所有可能占用 CPU 时间的不必要的应用程序和服务。
  • 确保您的孩子在运行基准测试时没有开始在您的 PC 上观看电影。硬件 MPEG 解码往往会影响结果。（自动播放让我两岁的孩子通过插入磁盘来启动 Despicable Me。是的。）

正如我在对@Paul R 的评论回复中所说，考虑使用 OpenCL，因为它可以让您轻松地在 GPU 和 CPU 上运行相同的代码，而无需重新实现它。

（回想起来，这些可能非常明显。）

于 2011-10-05T15:05:05.280 回答
4

作为参考，我用时间测量做了一个类似的例子。使用 GTX 660，GPU 加速为 24 倍，其中除了实际计算之外，它的操作还包括数据传输。

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>
#include <time.h>

#define N (1024*1024)
#define M (10000)
#define THREADS_PER_BLOCK 1024

void serial_add(double *a, double *b, double *c, int n, int m)
{
    for(int index=0;index<n;index++)
    {
        for(int j=0;j<m;j++)
        {
            c[index] = a[index]*a[index] + b[index]*b[index];
        }
    }
}

__global__ void vector_add(double *a, double *b, double *c)
{
    int index = blockIdx.x * blockDim.x + threadIdx.x;
        for(int j=0;j<M;j++)
        {
            c[index] = a[index]*a[index] + b[index]*b[index];
        }
}

int main()
{
    clock_t start,end;

    double *a, *b, *c;
    int size = N * sizeof( double );

    a = (double *)malloc( size );
    b = (double *)malloc( size );
    c = (double *)malloc( size );

    for( int i = 0; i < N; i++ )
    {
        a[i] = b[i] = i;
        c[i] = 0;
    }

    start = clock();
    serial_add(a, b, c, N, M);

    printf( "c[0] = %d\n",0,c[0] );
    printf( "c[%d] = %d\n",N-1, c[N-1] );

    end = clock();

    float time1 = ((float)(end-start))/CLOCKS_PER_SEC;
    printf("Serial: %f seconds\n",time1);

    start = clock();
    double *d_a, *d_b, *d_c;


    cudaMalloc( (void **) &d_a, size );
    cudaMalloc( (void **) &d_b, size );
    cudaMalloc( (void **) &d_c, size );


    cudaMemcpy( d_a, a, size, cudaMemcpyHostToDevice );
    cudaMemcpy( d_b, b, size, cudaMemcpyHostToDevice );

    vector_add<<< (N + (THREADS_PER_BLOCK-1)) / THREADS_PER_BLOCK, THREADS_PER_BLOCK >>>( d_a, d_b, d_c );

    cudaMemcpy( c, d_c, size, cudaMemcpyDeviceToHost );


    printf( "c[0] = %d\n",0,c[0] );
    printf( "c[%d] = %d\n",N-1, c[N-1] );


    free(a);
    free(b);
    free(c);
    cudaFree( d_a );
    cudaFree( d_b );
    cudaFree( d_c );

    end = clock();
    float time2 = ((float)(end-start))/CLOCKS_PER_SEC;
    printf("CUDA: %f seconds, Speedup: %f\n",time2, time1/time2);

    return 0;
}
于 2014-09-19T05:55:13.250 回答
2

我同意 David 关于 OpenCL 是测试这一点的好方法的评论，因为在 CPU 和 GPU 上运行的代码之间切换是多么容易。如果您能够在 Mac 上工作，Apple 提供了一些使用 OpenCL 进行 N 体模拟的示例代码，内核在 CPU、GPU 或两者上运行。您可以在它们之间实时切换，并在屏幕上显示 FPS 计数。

对于一个更简单的情况，他们有一个“hello world”OpenCL 命令行应用程序，它以类似于 David 所描述的方式计算平方。这很可能可以毫不费力地移植到非 Mac 平台上。要在 GPU 和 CPU 使用率之间切换，我相信您只需要更改

int gpu = 1;

在 hello.c 源文件中的行中，CPU 为 0，GPU 为 1。

Apple 在他们的主要 Mac 源代码列表中有更多的 OpenCL 示例代码。

David Gohara 博士在有关该主题的介绍性视频会议的最后（大约 34 分钟）执行分子动力学计算时，举了一个 OpenCL 的 GPU 加速示例。在他的计算中，他看到从在 8 个 CPU 内核上运行的并行实现到单个 GPU 的加速大约提高了 27 倍。同样，这不是最简单的示例，但它展示了一个真实世界的应用程序以及在 GPU 上运行某些计算的优势。

我还使用 OpenGL ES 着色器对移动空间进行了一些修改，以执行基本计算。我发现，在 GPU 上作为着色器运行时，在图像上运行的简单颜色阈值着色器比在 CPU 上为该特定设备执行的相同计算快大约 14-28 倍。

于 2011-10-05T20:21:46.650 回答