Документация

Элементно-мудрый математический пример

Этот пример показывает, как создать ядра CUDA из функций, которые содержат поэлементные математические операции. Предположим, что вы хотите задать квадрат каждому элементу массива матрицы x и шкалы в множителе 1/(i+j), где i,j - это индексы строка и столбец. Можно реализовать этот пример как функцию MATLAB.

function [y] = myFun(x)

y = zeros(size(x));
for i = 1:size(x,1)
    for j = 1:size(x,2)
        y(i,j)=(x(i,j)^2)/(i+j);
    end
end
end

Подготовка myFun к генерации кода

Первый оператор zeros(size(A)) в myFun функция предназначена для инициализации вектора результатов y в нули. Для генерации кода CUDA предварительно выделите память для y без возникновения накладных расходов на инициализацию памяти в нули. Замените эту линию на coder.nullcopy(zeros(size(y))).

Чтобы создать ядра CUDA из циклов, GPU Coder™ предоставляет другую прагму coder.gpu.kernel. Установка этой прагмы ядра переопределяет весь анализ параллельного цикла. Если вы не задаете никаких параметров, GPU Coder определяет границы ядра на основе границ цикла и размера входа. Он предоставляет способ для вас задать параметры запуска ядра, такие как thread и block размеры. Однако используйте его только, когда вы знаете, что цикл безопасен для параллелизации. Потому что myFun пример прост и не требует спецификации параметров запуска ядра, можно использовать coder.gpu.kernelfun прагма для генерации ядер CUDA.

С этими модификациями исходный myFun функция подходит для генерации кода.

function [y] = myFun(x) %#codegen

y = coder.nullcopy(zeros(size(x)));
coder.gpu.kernelfun();
for i = 1:size(x,1)
    for j = 1:size(x,2)
        y(i,j)=(x(i,j)^2)/(i+j);
    end
end
end

Сгенерированный код CUDA

Когда вы генерируете код CUDA с помощью приложения GPU Coder или из командной строки, GPU Coder создает одно ядро, которое выполняет операцию квадрирования и масштабирования. Ниже представлен фрагмент myFun_kernel1 код ядра.

static __global__ __launch_bounds__(512, 1) void myFun_kernel1(const real_T *x,
  real_T *y)
{
...
threadId = ((((gridDim.x * gridDim.y * blockIdx.z + gridDim.x * blockIdx.y) +
                blockIdx.x) * (blockDim.x * blockDim.y * blockDim.z) +
               threadIdx.z * blockDim.x * blockDim.y) + threadIdx.y * blockDim.x)
    + threadIdx.x;
  i = (int32_T)(threadId / 512U);
  j = (int32_T)(threadId - (uint32_T)i * 512U);
  if ((!(j <= 512)) && (!(i <= 512))) {
    y[i + (j << 9)] = x[i + (j << 9)] * x[i + (j << 9)] / ((real_T)(i + j) + 2.0);
  }
}

Ниже представлен фрагмент основного myFun функция. Перед вызовом myFun_kernel1, существует единственная cudaMemcpy вызов, который передает матрицу x от хоста (x) к устройству (gpu_x). Ядро имеет 512 блоков, содержащих 512 потоков на блок, согласующихся с размером входного вектора. Второй cudaMemcpy вызов копирует результат расчета обратно на хост.

cudaMemcpy((void *)gpu_x, (void *)x, 2097152ULL, cudaMemcpyHostToDevice);
myFun_kernel1<<<dim3(512U, 1U, 1U), dim3(512U, 1U, 1U)>>>(gpu_x, gpu_y);
cudaMemcpy((void *)y, (void *)gpu_y, 2097152ULL, cudaMemcpyDeviceToHost);

Ограничения