Умножение матриц. Простейшая реализация (пример)
Произведение двух квадратных матриц A и B размера N*N, N кратно 16
Матрицы расположены в глобальной памяти
По одной нити на каждый элемент произведения
2D блок - 16*16
2D грид
На каждый элемент
2*N арифметических операций
2*N обращений к глобальной памяти
Memory bound (медленный доступ к памяти)
#define BLOCK_SIZE 16
__global__ void matMult ( float * a, float * b, int n, float * c )
{
int bx = blockIdx.x;
int by = blockIdx.y;
int tx = threadIdx.x;
int ty = threadIdx.y;
float sum = 0.0f;
int ia = n * BLOCK_SIZE * by + n * ty;
int ib = BLOCK_SIZE * bx + tx;
int ic = n * BLOCK_SIZE * by + BLOCK_SIZE * bx;
for ( int k = 0; k < n; k++ )
sum += a [ia + k] * b [ib + k*n];
c [ic + n * ty + tx] = sum;
}
int numBytes = N * N * sizeof ( float );
float * adev, * bdev, * cdev ;
dim3 threads ( BLOCK_SIZE, BLOCK_SIZE );
dim3 blocks ( N / threads.x, N / threads.y);
cudaMalloc ( (void**)&adev, numBytes ); // allocate DRAM
cudaMalloc ( (void**)&bdev, numBytes ); // allocate DRAM
cudaMalloc ( (void**)&cdev, numBytes ); // allocate DRAM
// copy from CPU to DRAM
cudaMemcpy ( adev, a, numBytes, cudaMemcpyHostToDevice );
cudaMemcpy ( bdev, b, numBytes, cudaMemcpyHostToDevice );
matMult<<<blocks, threads>>> ( adev, bdev, N, cdev );
cudaThreadSynchronize();
cudaMemcpy ( c, cdev, numBytes, cudaMemcpyDeviceToHost );
// free GPU memory
cudaFree ( adev );
cudaFree ( bdev );
cudaFree ( cdev );
Общие правила по оформлению программ
Программа
- должна выполняться параллельно
- должна распределять память
- должна использовать векторные операции
Программа должна делать проверки на ошибки:
- Наличие аппаратной поддержки CUDA
- Открылся ли файл исходных данных
- Правилен ли формат данных?
Программа должна быть скомпилирована с опцией Release и запускаться на Windows 7 CUDA Toolkit 4.2
Программа должна быть скомпилирована CUDA 3.2, 4.2, OpenCl, OpenACC
Программа должна компилироваться. Для этого должен быть приложен vcproj для VS2008, VS2010 либо makefile + .bat
Варианты заданий (* отмечены задачи повышенной сложности)
- Асинхронно вывести на экран средствами видеокарты заготовленные текстуры
Сложение 1000000 матриц (сравнение производительности с CPU)
Перемножение матриц
Поиск А-1 для заданной матрицы
Обращение матрицы методом Монте-карло
Поиск решений F(x) = 0 на интервале
Метод наименьших квадратов
Гистограмма
Битоническая сортировка
Редукция
Thrust сортировка 1000000 матриц
Поразрядная сортировка
Найти сумму квадратов
Найти среднеквадратичное отклонение
Найти дисперсию
Вейвлет преобразование Haar’a
*Метод конечных элементов для уравнения Гельмгольца в заданной области.
Свертка с использованием FFT
FFT 3D
Решение краевых задач, системы линейных алгебраических уравнений
Решение 1D/2D задачи диффузии
Решение СЛАУ с треугольной разряженной матрицей
Решение уравнения Пуассона
Оптимизация алгоритма reduce
Оптимизация алгоритма scan
Найти сумму квадратов методом scan
Оптимизация алгоритма histogram
Аппроксимировать sin(x) кубическим сплайном
Генерация случайных чисел
Вывести на экран 10^8 чисел менее чем за 2 секунды.