Пример cuSOLVER

В этом примере решаются системы линейных уравнений Ax = B для x с помощью библиотеки cuSOLVER. Матрицы A и B должно иметь одинаковое количество строк. Если A является скаляром, то A\B эквивалентно A.\B. Если A является квадратной матрицей n-by-n и B является матрицей с n строками, то x = A\B является решением уравнения A*x = B, если он существует. Реализация MATLAB ® backslash является:

function [x] = backslash(A,b)
if (isscalar(A))
    x = coder.nullcopy(zeros(size(b)));
else
    x = coder.nullcopy(zeros(size(A,2),size(b,2)));
end

x = A\b;

end

Подготовиться `backslash` для создания ядра

GPU Coder™ не требует специальной прагматики для генерации вызовов библиотек. Как и раньше, существует два способа создания ядер CUDA ® -coder.gpu.kernelfun и coder.gpu.kernel. В этом примере мы используем coder.gpu.kernelfun pragma для создания ядер CUDA. Измененный backslash функция:

function [x] = backslash(A,b) %#codegen

if (isscalar(A))
    x = coder.nullcopy(zeros(size(b)));
else
    x = coder.nullcopy(zeros(size(A,2),size(b,2)));
end

coder.gpu.kernelfun()
x = A\b;

end

Примечание

Для замены математических операторов и функций на реализации библиотеки cuSOLVER требуется минимальный размер входных данных. Минимальное пороговое значение составляет 128 элементов.

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

При создании кода CUDA кодер графического процессора создает вызовы функций для инициализации библиотеки cuSOLVER, выполнения mldivide и освободить аппаратные ресурсы, используемые библиотекой cuSOLVER. Фрагмент сгенерированного кода CUDA:

  cusolverEnsureInitialization();

  /*    Copyright 2017 The MathWorks, Inc. */
  cudaMemcpy(b_gpu_A, A, 1152UL, cudaMemcpyHostToDevice);  
  blackslash_kernel1<<<dim3(1U, 1U, 1U), dim3(160U, 1U, 1U)>>>(b_gpu_A,gpu_A);
  cudaMemcpy(b_A, gpu_A, 1152UL, cudaMemcpyDeviceToHost);
  cusolverDnDgetrf_bufferSize(cusolverGlobalHandle, 12, 12, &gpu_A[0], 12,
    &cusolverWorkspaceReq);
  cusolverWorkspaceTypeSize = 8;
  cusolverInitWorkspace();
  cudaMemcpy(gpu_A, b_A, 1152UL, cudaMemcpyHostToDevice);
  cusolverDnDgetrf(cusolverGlobalHandle, 12, 12, &gpu_A[0], 12, (real_T *)
                   cusolverWorkspaceBuff, &gpu_ipiv_t[0], gpu_info_t);
  A_dirtyOnGpu = true;
  cudaMemcpy(&info_t, gpu_info_t, 4UL, cudaMemcpyDeviceToHost);

Для инициализации библиотеки cuSOLVER и создания дескриптора контекста библиотеки cuSOLVER используется функция cusolversEnsureInitialization() требования cusolverDnCreate() cuSOLVER API. Он выделяет аппаратные ресурсы на хосте и устройстве.

static void cusolverEnsureInitialization(void)
{
  if (cusolverGlobalHandle == NULL) {
    cusolverDnCreate(&cuSolverGlobalHandle);
  }
}

backslash_kernel1 нуль прокладок матрицы A. Это ядро запускается с одним блоком из 512 потоков.

static __global__ __launch_bounds__(160, 1) void backslash_kernel1(const real_T *
  A, real_T *b_A)
{
  int32_T threadId;
  ;
  ;
  threadId = (int32_T)(((gridDim.x * gridDim.y * blockIdx.z + gridDim.x *
    blockIdx.y) + blockIdx.x) * (blockDim.x * blockDim.y * blockDim.z) +
                       (int32_T)((threadIdx.z * blockDim.x * blockDim.y +
    threadIdx.y * blockDim.x) + threadIdx.x));
  if (!(threadId >= 144)) {
    /*    Copyright 2017 The MathWorks, Inc. */
    b_A[threadId] = A[threadId];
  }
}

Звонки в cudaMemcpy перенести матрицу, A от хоста к устройству. Функция cusolverDnDgetrf вычисляет факторизацию LU матрицы m × n:

P*A = L*U

где A - матрица m × n, P - перестановочная матрица, L - нижняя треугольная матрица с единичной диагональю, а U - верхняя треугольная матрица.

Автономный код cuSOLVER

Для таких функций, как qr , которые имеют только частичную поддержку в cuSOLVER, GPU Coder использует библиотеку LAPACK, где это необходимо. Для функций MEX генератор кода использует библиотеку LAPACK, входящую в состав MATLAB. Для автономного кода генератор кода использует указанную библиотеку LAPACK. Чтобы указать библиотеку LAPACK, выполните следующие действия.

В командной строке определите свою собственную coder.LAPACKCallback класс, содержащий информацию о библиотеке LAPACK, и назначить ее CustomLAPACKCallback свойства объекта конфигурации кода.
В приложении GPU Coder установите пользовательский обратный вызов библиотеки LAPACK в библиотеку LAPACK.

Например, для создания автономного исполняемого файла можно использовать следующий сценарий создания кода. Здесь myLAPACK является именем пользовательского coder.LAPACKCallback класс, содержащий информацию о библиотеке LAPACK.

cfg = coder.gpuConfig('exe');
cfg.CustomLAPACKCallback = 'myLAPACK';
cfg.GenerateExampleMain = 'GenerateCodeAndCompile';

classdef myLAPACK < coder.LAPACKCallback
    methods (Static)
        function hn = getHeaderFilename()
            hn = 'lapacke.h';
        end
        function updateBuildInfo(buildInfo, buildctx)
            [~,linkLibExt] = buildctx.getStdLibInfo();
            cudaPath = getenv('CUDA_PATH'); 
            libPath = 'lib\x64';            
            
            buildInfo.addIncludePaths(fullfile(cudaPath,'include'));
            libName = 'cusolver';
            libPath = fullfile(cudaPath,libPath);
            buildInfo.addLinkObjects([libName linkLibExt], libPath, ...
                '', true, true);
            
            lapackLocation = 'C:\LAPACK\win64'; % specify path to LAPACK libraries 
            
            includePath = fullfile(lapackLocation,'include');
            buildInfo.addIncludePaths(includePath);
            libPath = fullfile(lapackLocation,'lib');
            libName = 'mllapack';
            
            buildInfo.addLinkObjects([libName linkLibExt], libPath, ...
                '', true, true);
            buildInfo.addDefines('HAVE_LAPACK_CONFIG_H');
            buildInfo.addDefines('LAPACK_COMPLEX_STRUCTURE');
        end
    end
end

Дополнительные сведения см. в разделе Ускорение линейной алгебры в сгенерированном автономном коде с помощью вызовов LAPACK.

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