Тайм-аут в CUDA?/ Fermi / GTX465 - PullRequest
Новая
       5

Тайм-аут в CUDA?/ Fermi / GTX465

1 голос
/ 12 июля 2010

Я использую CUDA SDK 3.1 на MS VS2005 с графическим процессором GTX465 1 ГБ. У меня есть такая функция ядра: 

__global__ void CRT_GPU_2(float *A, float *X, float *Y, float *Z, float *pIntensity, float *firstTime, float *pointsNumber)
{


  int holo_x = blockIdx.x*20 + threadIdx.x;
  int holo_y = blockIdx.y*20 + threadIdx.y;

  float k=2.0f*3.14f/0.000000054f;

  if (firstTime[0]==1.0f)
  {
   pIntensity[holo_x+holo_y*MAX_FINAL_X]=0.0f; 
  }

  for (int i=0; i<pointsNumber[0]; i++)
  {
   pIntensity[holo_x+holo_y*MAX_FINAL_X]=pIntensity[holo_x+holo_y*MAX_FINAL_X]+A[i]*cosf(k*sqrtf(pow(holo_x-X[i],2.0f)+pow(holo_y-Y[i],2.0f)+pow(Z[i],2.0f)));
  }

  __syncthreads(); 


}

и это функция, которая вызывает функцию ядра: 

extern "C" void go2(float *pDATA, float *X, float *Y, float *Z, float *pIntensity, float *firstTime, float *pointsNumber)
{
 dim3 blockGridRows(MAX_FINAL_X/20,MAX_FINAL_Y/20);
 dim3 threadBlockRows(20, 20);

 CRT_GPU_2<<<blockGridRows, threadBlockRows>>>(pDATA, X, Y, Z, pIntensity,firstTime, pointsNumber); 
 CUT_CHECK_ERROR("multiplyNumbersGPU() execution failed\n");
 CUDA_SAFE_CALL( cudaThreadSynchronize() );
}

Я загружаю в цикле все параметры этой функции (например, 4096 элементов для каждого параметра в одной итерации цикла). Всего я хочу сделать это ядро ​​для 32768 элементов для каждого параметра после всех итераций цикла.

MAX_FINAL_X - 1920, а MAX_FINAL_Y - 1080.

Когда я запускаю alghoritm, первая итерация идет очень быстро, и после еще одной или двух итераций я получаю информацию об ошибке тайм-аута CUDA. Я использовал этот алгоритм на GPU GTX260, и, насколько я помню, он работал лучше ...

Не могли бы вы помочь мне ... может быть, я делаю какую-то ошибку в соответствии с новой аркой Ферми в этом алгоритме?

Ответы [ 3 ]

1 голос
/ 12 июля 2010
  1. Будет лучше позвонить CUT_CHECK_ERROR после cudaThreadSynchronize().Поскольку ядро ​​работает асинхронно, и вы должны дождаться окончания ядра, чтобы узнать об ошибках ... Возможно, во второй итерации вы получите ошибку от первого использования ядра.
  2. Убедитесь, что в наиболее интересной переменной pointsNumber[0] указан верный номер (это может вызвать длинный внутренний цикл).
  3. Вы также можете улучшить скорость работы ядра:
    • Используйте лучшие блоки.Конфигурация потоков 20x20 приведет к очень медленному использованию памяти (см. Руководство по программированию и Рекомендации).Попробуйте использовать блоки 16х16.
    • Не использовать функцию pow(..., 2.0).Макрос SQR быстрее использовать (#define SQR(x) (x)*(x))
    • Вы не используете общий мем, поэтому __syncthreads() не требуется.

PS:Вы также можете передавать значения параметров в функции CUDA, а не только указатели.Скорость будет одинаковой.

PPS: пожалуйста, улучшите читабельность кода ... Теперь вы должны отредактировать шесть мест, чтобы изменить конфигурацию блока ... Внутри ядра вы можете использовать переменную blockDim и использовать константыфункция go2.Вы также можете использовать bool firstTime - это будет НАМНОГО лучше, чем float.

1 голос
/ 14 апреля 2012

В цикле ядра вы пишете в том же массиве, из которого читаете - для глобального использования памяти это хуже всего, потому что деформации из разных блоков ждут друг друга.

1 голос
/ 12 июля 2010

Ваш графический процессор подключен к дисплею?Если это так, я считаю, что по умолчанию выполнение ядра будет прервано через 5 секунд.Вы можете проверить время ожидания выполнения ядра, используя cudaGetDeviceProperties - см. справочную страницу

...