Программирование на CUDA Урок 1. Hello World (Сложение векторов)

Программа "Hello world" при изучении языка программирования CUDA C носит
нестандартный вид и представляет собой сложение двух векторов. В дальнейшем на этом же примере мы рассмотрим преимущество графического процессора относительно центрального в хорошо распараллеливаемых задачах.

В целом язык CUDA C представляет собой расширение языка C и будет понятен программистам, знакомым с Си-подобными языками.

Прежде чем перейти к написанию первой программы стоит рассмотреть основные элементы языка.

При программировании на CUDA C разделяют основной процессор с оперативной памятью — host, и графический процессор с памятью (видеокарта, GPU) — device.
Обратимся к коду программы для сложения векторов, поставляемой вместе с CUDA Toolkit (в нашем случае 8.0), но с русифицированными комментариями:

В начале программы мы видим объявление функции

cudaError_t addWithCuda(int *c, const int *a, const int *b, unsigned int size);

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

Начнем рассмотрение примера с середины: с функции__global__ void addKernel(int *c, const int *a, const int *b)

эта функция имеет модификатор __global__, который говорит о том, что эта функция будет исполняться на GPU (device) и эта функция будет обрабатываться компилятором CUDA (остальные же функции будут обрабатываться обычным компилятором для CPU).

Рассматривая содержимая этой функции мы встретим незнакомую инструкцию:

int i = threadIdx.x;

Как вы уже догадались, GPU и в частности CUDA устройства являются многопоточными, но в отличии от многоядерных CPU, CUDA строится по принципу SIMT (одна инструкция и несколько потоков). А структура threadIdx является местом хранения данных о конкретном потоке (нити), а команда представленная выше присваивает переменной i значение, равное номеру потока(нити). И далее нить работает со "своими" данными из массива и эта работа должна быть вам понятна. 

Перейдем к функции main в ней не происходит ничего необычного для C/C++ программиста, кроме разве что:

	cudaStatus = cudaDeviceReset();
	if (cudaStatus != cudaSuccess) {
	fprintf(stderr, "cudaDeviceReset failed!");
	return 1;
	}

Данный команды сбрасывают состояние устройств перед выходом для корректной работы инструментов профилирования и отслеживания.

Теперь перейдем к основной части работы с GPU — к функции-обертке 

cudaError_t addWithCuda(int *c, const int *a, const int *b, unsigned int size);

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

cudaStatus = cudaSetDevice(0);

Зачастую при работе с данными на устройстве имеются несколько GPU и данная команда выбирает первое из них, в целом, как мы рассмотрим позднее можно получить список GPU и выбрать предпочтительное.

cudaStatus = cudaMalloc((void**)&dev_c, size * sizeof(int));

Выделение памяти на устройстве — к этой памяти нельзя напрямую обращаться из основного кода. И попытки прямого копирования или обработки этих данных приведут к ошибке выполнения программи.

cudaStatus = cudaMemcpy(dev_a, a, size * sizeof(int), cudaMemcpyHostToDevice);

Копирование данных — достаточно долгая, но необходимая процедура, т.к. GPU не может напрямую работать с данными в оперативной памяти, да и будет это все равно дольше.

Аргумент cudaMemcpyHostToDevice говорит о направлении копирования из Host в Device. Для обратного копирования в оперативную память используется константа cudaMemcpyDeviceToHost.

Далее самая главная команда ради которой ранее все и делалось.

addKernel<<<1, size>>>(dev_c, dev_a, dev_b);

Она запускает функцию addKernel с 1 блоком и количеством потоков-нитей равным size.

Предпоследняя команда, которую мы не рассмотрели:

cudaStatus = cudaDeviceSynchronize();

Как и сказано в комментарии, ждут пока все блоки и нити завершат свою работу. Фактически является барьерной синхронизацией в конце GPU кода.

Последняя и "самая главная" команда для программиста:

cudaFree(dev_c);

освобождение памяти, про нее нельзя забывать, иначе пользователь не скажет Вам спасибо за мусор в памяти GPU (наличие которого еще и отследить то не так просто, как в оперативной памяти).

Ссылки:

Nvidia Developer

Загрузка CUDA Toolkit