coder.gpu.constantMemory

Прагма, которая преобразует переменную в постоянную память на графическом процессоре

Синтаксис

Описание

пример

coder.gpu.constantMemory(v) отображает переменную v в постоянное пространство памяти на графическом процессоре. Поместите эту прагму в параллелизируемый цикл. Если GPU Coder™ генерирует ядро для цикла, он загружается v в переменную постоянной памяти устройства. Он заменяет любой доступ к этой переменной в ядре доступом к переменной постоянной памяти. В ядре переменная v должно быть доступно только для чтения. В противном случае GPU Coder игнорирует эту прагму. Используйте эту прагму, когда каждый поток обращается к каждому элементу массива параметров или матрицы.

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

Примеры

свернуть все

В этом примере показано, как сопоставить вход в постоянное пространство памяти на графическом процессоре с помощью coder.gpu.constantMemory прагма.

Написание функции точки входа myFun который принимает два входов a размера 256x256 и постоянные k размера 1x3. Функция имеет вложенное for-циклы, который добавляет константы к каждому элементу a. Чтобы создать ядро, разместите coder.gpu.kernel() прагма вне вложенного for-цикл. coder.gpu.constantMemory(k) помещает вход только для чтения k в постоянную память графический процессор.

function b = myFun(a,k)
  b = coder.nullcopy(zeros(size(a)));
  coder.gpu.kernel();
    for j = 1:256
      for i = 1:256
        coder.gpu.constantMemory(k);  
        b(i,j) = a(i,j) + k(1) + k(2) + k(3);
      end
    end
end

Создайте объект строения для генерации кода MEX.

cfg = coder.gpuConfig('mex');

Задайте массив ячеек input который объявляет размер и тип данных входов a,k в myFun функций.

input = {ones(256),ones(1,3)}

Сгенерируйте MEX-функцию myFun_mex при помощи -config, -args, и -report опции, чтобы задать строение, предоставить входные параметры и сгенерировать отчет генерации кода.

codegen -config cfg -args input -report myFun

В отчете на вкладке C code нажмите myFun.cu.

Переменная только для чтения k объявлен как const_k при помощи __constant__ как показано на фрагменте кода.

/* Variable Definitions */
__constant__ real_T const_k[3];

cudaMemcpyToSymbol вызов копирует значение k от хоста к устройству постоянной памяти const_k.

  cudaMemcpyToSymbol(const_k, k, 24U, 0U, cudaMemcpyHostToDevice);
  cudaMemcpy(gpu_a, a, 524288U, cudaMemcpyHostToDevice);
  myFun_kernel1<<<dim3(128U, 1U, 1U), dim3(512U, 1U, 1U)>>>(gpu_a, gpu_b);
  cudaMemcpy(b, gpu_b, 524288U, cudaMemcpyDeviceToHost);  

Тело ядра обращается к постоянной const_k и добавляет его к каждому элементу a

static __global__ __launch_bounds__(512, 1) void myFun_kernel1(const real_T *a,
 real_T *b)
{
  int32_T i;
  int32_T j;
  int32_T threadIdX;
  threadIdX = (int32_T)(blockDim.x * blockIdx.x + threadIdx.x);
  i = threadIdX / 256;
  j = threadIdX - i * 256;
  if ((!(j >= 256)) && (!(i >= 256))) {
    b[i + (j << 8)] = ((a[i + (j << 8)] + const_k[0]) + const_k[1]) + const_k[2];
  }
}

Входные параметры

свернуть все

Имя переменной, которая должна быть сопоставлена с пространством постоянной памяти на графическом процессоре.

Введенный в R2017b