Блог

Запуск ядра в 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;
  

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

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

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

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

    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;