coder.gpu.nokernel

Прагма, чтобы отключить циклы for создания ядра

Синтаксис

Описание

пример

coder.gpu.nokernel() прагма уровня цикла, что, когда помещено сразу, прежде чем цикл for препятствует тому, чтобы генератор кода генерировал ядра CUDA® для операторов в цикле. Эта прагма не требует никаких входных параметров.

Эта функция является функцией генерации кода. Это не оказывает влияния в MATLAB®.

Примеры

свернуть все

В этом примере показано, как использовать nokernel прагма в функции и препятствует тому, чтобы генератор кода генерировал ядра CUDA для операторов в цикле

В одном файле запишите функцию точки входа nestedLoop это признает, что два вектора вводят A,B из размера 32x512. Функция имеет два, вложил for- циклы различных длин итерации, один для работы вдоль столбца и один для работы вдоль строки. Первый вложенный цикл вычисляет сумму двух векторных входных параметров, в то время как второй вложенный цикл масштабирует сумму на коэффициент три.

function [C] = nestedLoop(A, B)
    G = zeros(32, 512);
    C = zeros(32, 512);        
   
    coder.gpu.kernelfun();
    % This nested loop will be fused
    for i = 1:32
       for j = 1:512
           G(i,j) = A(1,j) + B(1,j);
       end
    end

    coder.gpu.nokernel();  
    for i = 1:32
       for j = 1:512
           C(i,j) = G(i,j) * 3;
       end
    end    
end

Используйте codegen функция, чтобы сгенерировать MEX-функцию CUDA.

cfg = coder.gpuConfig('mex');
cfg.GenerateReport = true;
codegen -config cfg -args {ones(1,512,'double'),ones(1,512,'double')} nestedLoop

GPU Coder создает два ядра: nestedLoop_kernel1 выполнять расчет G(i,j) = A(1,j) + B(1,j); из первого вложенного цикла и nestedLoop_kernel2 ядро, чтобы выполнить расчет C(i,j) = G(i,j) * 3; из второго вложенного цикла. Второе ядро создается для внутреннего цикла второго вложенного цикла. noKernel прагма применима только к циклу, который сразу следует за оператором. Отрывки сгенерированных ядер показывают.

static __global__ __launch_bounds__(512, 1) void nestedLoop_kernel1(const real_T
  B[512], const real_T A[512], real_T G[16384])
{
  uint32_T threadId;
  ...
  if (i < 32) {
    G[i + (j << 5)] = A[j] + B[j];
  }
}
static __global__ __launch_bounds__(512, 1) void nestedLoop_kernel2(real_T G
  [16384], int32_T i, real_T C[16384])
{
  uint32_T threadId;
  ...;
  if (j < 512) {
    C[i + (j << 5)] = G[i + (j << 5)] * 3.0;
  }

Отрывок основной функции показывает, что генератор кода плавил первый вложенный цикл, как обозначено параметрами запуска ядра. Как отмечалось ранее, внешний контур второго вложенного цикла является тем, который не сопоставлен с ядром. Следовательно генератор кода помещает for-loop оператор незадолго до вызова второго ядра CUDA nestedLoop_kernel2.

void nestedLoop(const real_T A[512], const real_T B[512], real_T C[16384])
{
  int32_T i;
  ...
  //  These two loops will be fused
  cudaMemcpy(gpu_B, (void *)&B[0], 4096UL, cudaMemcpyHostToDevice);
  cudaMemcpy(gpu_A, (void *)&A[0], 4096UL, cudaMemcpyHostToDevice);
  nestedLoop_kernel1<<<dim3(32U, 1U, 1U), dim3(512U, 1U, 1U)>>>(*gpu_B, *gpu_A, *
    gpu_G);
  for (i = 0; i < 32; i++) {
    nestedLoop_kernel2<<<dim3(1U, 1U, 1U), dim3(512U, 1U, 1U)>>>(*gpu_G, i,
      *gpu_C);
    C_dirtyOnGpu = true;
  }
...
  cudaFree(*gpu_C);
}
Введенный в R2019a
Для просмотра документации необходимо авторизоваться на сайте