Timing Kernel запускается в CUDA при использовании Thrust - PullRequest
Новая
       21

Timing Kernel запускается в CUDA при использовании Thrust

3 голосов
/ 11 ноября 2011

Запуски ядра в CUDA обычно асинхронны, что (как я понимаю) означает, что после запуска ядра CUDA управление немедленно возвращается к ЦП. Процессор продолжает выполнять полезную работу, пока графический процессор занят если процессор не будет принудительно остановлен с помощью cudaThreadsynchronize() или cudaMemcpy().

Теперь я только начал использовать библиотеку Thrust для CUDA. Являются ли вызовы функций в Thrust? синхронный или асинхронный?

Другими словами, если я вызову thrust::sort(D.begin(),D.end());, где D - вектор устройства, имеет ли смысл измерять время сортировки, используя 

        start = clock();//Start

             thrust::sort(D.begin(),D.end());

        diff = ( clock() - start ) / (double)CLOCKS_PER_SEC;
        std::cout << "\nDevice Time taken is: " <<diff<<std::endl;

Если вызов функции асинхронный, тогда diff будет 0 секунд для любого вектора (что нежелательно для таймингов), но если он синхронный, я действительно получу производительность в реальном времени.

1 Ответ

4 голосов
/ 11 ноября 2011

Вызовы Thrust, которые вызывают ядра, являются асинхронными, как и базовые API-интерфейсы CUDA.Вызовы Thrust, которые копируют данные, являются синхронными, точно так же, как используются базовые API-интерфейсы CUDA.

Таким образом, ваш пример будет измерять только издержки при запуске ядра и установке на стороне хоста, а не саму операцию.Для определения времени вы можете обойти это, вызвав либо cudaThreadSynchronize, либо cudaDeviceSynchronize (позже в CUDA 4.0 или позже) после запуска ядра тяги.В качестве альтернативы, если вы включите операцию копирования после запуска ядра и запишете время остановки после этого, ваше время будет включать настройку, выполнение и время копирования.

В вашем примере это будет выглядеть примерно так: 

   start = clock();//Start 

   thrust::sort(D.begin(),D.end()); 
   cudaThreadSynchronize(); // block until kernel is finished

   diff = ( clock() - start ) / (double)CLOCKS_PER_SEC; 
   std::cout << "\nDevice Time taken is: " <<diff<<std::endl;
...