Простейший возможный пример, показывающий, как GPU превосходит CPU с помощью CUDA - PullRequest
25 голосов
/ 05 октября 2011

Я ищу максимально сжатый объем кода, который может быть закодирован как для CPU (с использованием g ++), так и для GPU (с использованием nvcc), для которого GPU постоянно превосходит CPU. Любой тип алгоритма является приемлемым.

Чтобы уточнить: я буквально ищу два коротких блока кода, один для процессора (использующий C ++ в g ++) и один для графического процессора (использующего C ++ в nvcc), для которого графический процессор превосходит. Предпочтительно в масштабе секунд или миллисекунд. Самая короткая возможная пара кодов.

Ответы [ 4 ]

37 голосов
/ 05 октября 2011

Прежде всего, я повторю свой комментарий: графические процессоры имеют высокую пропускную способность и большую задержку. Попытка заставить GPU превзойти CPU для выполнения наносекундной работы (или даже миллисекунды или второй работы) полностью упускает смысл делать GPU. Ниже приведен простой код, но для того, чтобы по-настоящему оценить преимущества производительности графического процессора, вам потребуется большой размер проблемы, чтобы амортизировать затраты на запуск по сравнению с ... в противном случае это бессмысленно. Я могу обыграть Ferrari в двухфутовой гонке, просто потому, что нужно повернуть ключ, запустить двигатель и нажать педаль. Это не значит, что я быстрее, чем Ferrari, каким-либо значимым образом.

Используйте что-то подобное в 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.

3 голосов
/ 05 октября 2011

Очень, очень простой способ - вычислить квадраты, скажем, для первых 100 000 целых чисел или для операции с большой матрицей.Его легко реализовать и использовать сильные стороны графических процессоров, избегая ветвления, не требуя стека и т. Д. Я сделал это с OpenCL против C ++ некоторое время назад и получил довольно удивительные результаты.(2-гигабайтный GTX460 показал примерно 40-кратную производительность двухъядерного основного процессора.)

Вы ищете пример кода или просто идеи?

Редактировать

40-кратный был двухъядерным, а не четырехъядерным процессором.

Некоторые указатели:

  • Убедитесь, что выне запускайте, скажем, Crysis, пока выполняете тесты.
  • Удалите все ненужные приложения и службы, которые могут похищать процессорное время.
  • Убедитесь, что ваш ребенок не начинает смотреть фильм наваш компьютер во время выполнения тестов.Аппаратное декодирование MPEG имеет тенденцию влиять на результат.(Автозапуск позволил моему двухлетнему ребенку запустить «Гадкого я», вставив диск. Да.)

Как я уже сказал в своем комментарии к @Paul R, рассмотрите возможность использования OpenCL, поскольку он легко позволит вам запуститьодин и тот же код на GPU и CPU без необходимости его повторной реализации.

(Это, вероятно, довольно очевидно в ретроспективе.)

2 голосов
/ 19 сентября 2014

Для справки я сделал аналогичный пример с измерениями времени. В GTX 660 ускорение графического процессора составило 24X, где его работа включает в себя передачу данных в дополнение к фактическим вычислениям.

#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;
} 
2 голосов
/ 06 октября 2011

Я согласен с комментариями Дэвида о том, что OpenCL - отличный способ проверить это, потому что легко переключаться между выполнением кода на CPU и GPU. Если вы можете работать на Mac, у Apple есть хороший пример кода, который выполняет моделирование N-тела с использованием OpenCL с ядрами, работающими на CPU, GPU или обоих. Вы можете переключаться между ними в режиме реального времени, и количество кадров в секунду отображается на экране.

Для гораздо более простого случая у них есть приложение командной строки OpenCL * hello world , которое вычисляет квадраты способом, аналогичным описанному Дэвидом. Это, вероятно, может быть перенесено на не-Mac платформы без особых усилий. Я полагаю, что для переключения между использованием графического процессора и процессора вам просто нужно изменить

int gpu = 1;

строка в исходном файле hello.c до 0 для CPU, 1 для GPU.

Apple имеет еще несколько примеров кода OpenCL в своем основном списке исходных кодов Mac .

Dr. У Дэвида Гохары был пример ускорения графического процессора в OpenCL при выполнении расчетов по молекулярной динамике в самом конце этого вводного видео-сеанса на тему (около минуты 34). В своих расчетах он видит примерно 27-кратное ускорение, перейдя от параллельной реализации, работающей на 8 процессорных ядрах, к одному графическому процессору. Опять же, это не самый простой пример, но он показывает реальное приложение и преимущество выполнения определенных вычислений на GPU.

Я также проделал некоторую работу в мобильном пространстве, используя шейдеры OpenGL ES для выполнения элементарных вычислений . Я обнаружил, что при использовании шейдера в графическом процессоре простой шейдер цветовой привязки по изображению был примерно в 14-28 раз быстрее, чем те же вычисления, выполняемые для этого конкретного устройства.

...