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

Пример элементной математики

В этом примере показано, как создавать ядра 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. Указание этой pragma ядра переопределяет весь анализ с параллельным циклом. Если параметры не указаны, кодер графического процессора определяет границы ядра на основе границ цикла и размера входных данных. Он позволяет задать параметры запуска ядра, такие как размеры потоков и блоков. Однако используйте его только в том случае, если известно, что цикл безопасен для параллелизма. Потому что myFun пример прост и не требует спецификации параметров запуска ядра, вы можете использовать coder.gpu.kernelfun pragma для создания ядер 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);

Ограничения