Программирование для GPU с использованием NVidia CUDA.

Программирование для GPU с использованием NVidia CUDA. Половинкин А.Н.

Содержание • Постановка задачи • Алгоритм вычисления функции axpy на GPU • Программная реализация

Постановка задачи • y := alpha*x + y • x, y – векторы • размерность – n • alpha – скаляр

Алгоритм вычисления функции ?axpy на GPU ysub,i n ... ... BLOCK_SIZE BLOCK_SIZE BLOCK_SIZE BLOCK_SIZE THREAD BLOCK 1 THREAD BLOCK 2 THREAD BLOCK i THREAD BLOCK K Каждый блок потоков занимается вычислением одного подвектораysub,iвектора y Каждый поток внутри блока потоков занимается вычислением одного элементаподвектораysub,i

Структура проекта • axpy.h– содержит определения (через define) размера блока и размеров матриц • axpy_gold.cpp • computeGold • axpy.cu • main • randomInit • printDiff • runAxpy • axpy_kernel.cu • axpy (kernel)

Реализация функции runAxpy (1) • void runAxpy(intargc, char** argv) • инициализируем устройство (device) CUT_DEVICE_INIT(argc, argv); • выделяем память на хосте для хранения векторов xи y unsigned intmem_size = sizeof(float) * n; float* h_x = (float*)malloc(mem_size); float* h_y = (float*)malloc(mem_size);

Реализация функции runAxpy (2) • инициализируем векторы x и y случайными значениями randomInit(h_x, n); randomInit(h_y, n); • выделяем память под векторы x и y на устройстве, копируем данные с хоста на устройство float* d_x; CUDA_SAFE_CALL(cudaMalloc((void**)&d_x, mem_size)); float* d_y; CUDA_SAFE_CALL(cudaMalloc((void**)&d_y, mem_size)); CUDA_SAFE_CALL(cudaMemcpy(d_x, h_x, mem_size, cudaMemcpyHostToDevice) ); CUDA_SAFE_CALL(cudaMemcpy(d_y, h_y, mem_size, cudaMemcpyHostToDevice) );

Реализация функции runAxpy (3) • создаем и инициализируем таймер unsigned int timer = 0; CUT_SAFE_CALL(cutCreateTimer(&timer)); CUT_SAFE_CALL(cutStartTimer(timer)); • определяем конфигурацию выполнения ядра (размер решетки блоков потоков и блока потоков) dim3 threads(BLOCK_SIZE); dim3 grid(n / threads.x); • запускаем ядро • копируем вычисленный вектор y с устройства на хост

Реализация функции runAxpy (4) • останавливаем таймер, выводим время вычислений, освобождаем ресурсы таймера CUT_SAFE_CALL(cutStopTimer(timer)); printf("Processing time: %f (ms) \n", cutGetTimerValue(timer)); CUT_SAFE_CALL(cutDeleteTimer(timer)); • вычисляем то же самое произведение на CPU float* reference = (float*) malloc(mem_size); computeGold(reference, alpha, h_x, h_y, n);

Реализация функции runMultiplication (5) • сравниваем результат, полученный на GPU, с результатом, полученным на CPU (по евклидовой норме) CUTBoolean res = cutCompareL2fe(reference, h_y, n, 1e-6f); printf("Test %s \n", (1 == res) ? "PASSED" : "FAILED"); if (res!=1) printDiff(reference, h_y, n); • освобождаем память

Реализация функции-ядра (1) • __global__ voidaxpy( int n, float alpha, float* x, float* y) • вычисляем координату текущего блока потоков и сохраняем её в переменную bid int bid = blockIdx.x • вычисляем координату текущего потока в блоке потоков и сохраняем её в переменную tid • вычисляем индекс элемента в исходном массиве, который будет обрабатываться текущим потоком int index = bid * BLOCK_SIZE + tid

Реализация функции-ядра (2) • вычисляем значение элемента массива, обрабатываемого текущим потоком y[index] = alpha * x[index] + y[index]

Литература • Nvidia CUDA Programming Guide • Многочисленные курсы по CUDA: • http://courses.ece.uiuc.edu/ece498/al1/Syllabus.html • http://www.nvidia.ru/object/cuda_state_university_courses_ru.html(на русском языке)

Вопросы ?

Программирование для GPU с использованием NVidia CUDA.