Сроки ядра CUDA

Question

привет, каждый из них в настоящее время работает над определением некоторых из моего кода CUDA. Я был в состоянии использовать их, используя события. Мое ядро ​​работало в течение 19 мс. Как-то я нахожу это сомнительным, потому что, когда я запускал последовательную реализацию этого, он составлял около 5000 мс. Я знаю, что код должен работать быстрее, но должно ли это так быстро?

Я использую функции обертки для вызова ядер куды в моей программе cpp. Должен ли я называть их там или в файле .cu? Благодаря!

Answer 1

В моем собственном коде я использую функцию clock(), чтобы получить точные тайминги. Для удобства, я макросы

enum { 
    tid_this = 0, 
    tid_that, 
    tid_count 
    }; 
__device__ float cuda_timers[ tid_count ]; 
#ifdef USETIMERS 
#define TIMER_TIC clock_t tic; if (threadIdx.x == 0) tic = clock(); 
#define TIMER_TOC(tid) clock_t toc = clock(); if (threadIdx.x == 0) atomicAdd(&cuda_timers[tid] , (toc > tic) ? (toc - tic) : (toc + (0xffffffff - tic))); 
#else 
#define TIMER_TIC 
#define TIMER_TOC(tid) 
#endif 

Они могут быть использованы для инструмента код устройства следующим образом:

__global__ mykernel (...) { 

    /* Start the timer. */ 
    TIMER_TIC 

    /* Do stuff. */ 
    ... 

    /* Stop the timer and store the results to the "timer_this" counter. */ 
    TIMER_TOC(tid_this); 

    } 

Вы можете прочитать cuda_timers в принимающем коде.

Несколько замечаний:

• таймеры работают на поблочно основе, то есть, если у вас есть 100 блоков исполняющих то же ядро, сумма всех их времени будет храниться.
• Таймеры подсчитывают количество тактов. Чтобы получить количество миллисекунд, разделите его на количество ГГц на вашем устройстве и умножьте на 1000.
• Таймеры могут немного замедлить ваш код, поэтому я их обернул в #ifdef USETIMERS, чтобы вы могли отключить их без труда.
• Хотя clock() возвращает целочисленные значения типа clock_t, я сохраняю накопленные значения как float, в противном случае значения будут обертываться для ядер, которые занимают больше нескольких секунд (накопленные по всем блокам).
• Выбор (toc > tic) ? (toc - tic) : (toc + (0xffffffff - tic))) необходим в случае, когда счетчик часов обходит вокруг.

Answer 2

Очевидным способом проверить, работает ли ваша программа, является сравнение вывода с результатами реализации на основе ЦП. Если вы получаете тот же результат, он работает по определению, правильно? :)

Если ваша программа экспериментальна таким образом, что она на самом деле не производит какой-либо проверяемый вывод, есть хороший шанс, что компилятор оптимизировал некоторые (или все) вашего кода. Компилятор удалит код, который не вносит вклад в выходные данные. Это может привести, например, к тому, что все содержимое ядра будет удалено, если последний оператор, который хранит вычисленное значение, закомментирован.

Что касается вашего ускорения. 5000 мс/19 мс = 263x, что маловероятно, даже для алгоритмов, идеально подходящих для архитектуры графического процессора.

Answer 3

Хорошо, если вы написали свой код CUDA правильно, да, это может быть намного быстрее. Думаю об этом. Вы переместили код из последовательного исполнения на одном процессоре на параллельное выполнение на сотнях процессоров, в зависимости от вашей модели графического процессора. Моя средняя карта за 179 долларов США имеет 480 ядер. Некоторые из доступных теперь имеют 1500 ядер. Очень возможно получить 100-кратные прыжки с CUDA, особенно если ваше ядро ​​намного больше вычисляется, чем связано с памятью.

Тем не менее, убедитесь, что вы измеряете то, что, по вашему мнению, вы измеряете.Если вы вызываете свое ядро ​​CUDA без использования каких-либо явных потоков, тогда вызов синхронен с хостом, и ваши тайминги должны быть точными. Если вы вызываете свое ядро ​​с помощью потока, вам нужно вызвать cudaDeviceSynchronise() или оживить код хоста в событии, о котором сообщает ядро. Вызовы ядра, вызываемые в потоке, выполняются асинхронно с потоком хоста, поэтому измерения времени в потоке хоста не будут правильно отражать время ядра, если вы не сделаете поток хоста до тех пор, пока вызов ядра не будет завершен. Вы также можете использовать события CUDA для измерения прошедшего времени на графическом процессоре в пределах данного потока. См. Раздел 5.1.2 Руководства по лучшей практике CUDA в NVIDIA GPU Computing SDK 4.2.