Тест отрицательного индекса массива CUDA - PullRequest
Новая
       109

Тест отрицательного индекса массива CUDA

1 голос
/ 13 апреля 2020

Поскольку в C возможно индексировать отрицательное местоположение массива и go вне границ массива, этот код компилируется и «работает». 

__global__ void do_something_bad(int * in_a){
   in_a[-1] = 666; // assign a value to an out of bounds memory location
}

Я предполагаю, что приведенный выше код выполняет следующее (пожалуйста, дайте мне знать, если это предположение неверно): 

GPU memory before:
[0x00 = usually unused memory][0x01= Start of in_a][0x02 = in_a] ....
GPU memory after:
[0x00 = 666][0x01= Start of in_a][0x02 = in_a] ....

Таким образом, в памяти перед значением массива in_a устанавливается значение. Эта память до in_a могла содержать другие важные данные, но когда я тестирую ее, она не содержит ничего важного и, таким образом, не дает мне ошибки или провала теста.

К вашему сведению: я использую pycuda и тестирую мой код как go.

Я стараюсь не создавать молчаливых непредсказуемых ошибок в результате вышеизложенного. Конечно, в реальном случае пример -1 был бы рассчитан, и я упростил код до той проблемы, которую хочу решить.

Как определить эту ошибку и вызвать обнаруживаемую проблему, которую тестирует мой модуль можно забрать?

1 Ответ

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

Чтобы избежать тихой ошибки в ядре, я бы использовал утверждение , если вы не используете MacOS. Примерно так: 

#include <assert.h>

__global__ void do_something_bad(int* in_a){
  int indx;
  indx = 0; // A valid index
  assert(indx >= 0); // Lets the kernel continue
  in_a[indx] = 666;
  indx = -1; // An invalid index
  assert(indx >= 0); // Sends an error to stderr 
  in_a[indx] = 666; // This never gets executed  
}

int main(){
  int *a;
  cudaMalloc((void **)&a, 10*sizeof(int));
  do_something_bad<<<1,1>>>(a);
  cudaDeviceSynchronize();
}

Однако это может повлиять на производительность вашей программы. Из Руководства по программированию:

Утверждения для целей отладки. Они могут повлиять на производительность, поэтому рекомендуется отключить их в рабочем коде. Их можно отключить во время компиляции, определив макрос препроцессора NDEBUG перед включением assert.h

...