Exponenta Banner
exponenta event banner

gpucoder.reduce

Оптимизированная реализация графического процессора для операций сокращения

свернуть все на странице

Синтаксис

S = gpucoder.reduce (A, FUN)
S = gpucoder.reduce (A, {@ FUN1, @ FUN2,...})

Описание

пример

S = gpucoder.reduce(A,FUN) агрегирует значения, присутствующие во входном массиве A к одному значению с помощью данного дескриптора функции FUN. Продукция S является скаляром.

S = gpucoder.reduce(A,{@FUN1,@FUN2,...}) принимает входной массив и массив ячеек дескрипторов функций. Он агрегирует значения, присутствующие во входном массиве, в одно значение для каждого дескриптора функции, предоставленного в массиве ячеек. Размер выходного сигнала равен 1-by-N, где N - количество дескрипторов функций.

Генератор кода использует shuffle особенности для выполнения эффективного уменьшения на GPU. Несколько дескрипторов функций агрегируются внутри одного ядра на GPU.

Примеры

свернуть все

В этом примере создается код CUDA ® для поиска суммы и максимума элементов массива.

В одном файле запишите функцию точки входа multireduce который принимает матричный ввод A. Используйте gpucoder.reduce выполнение двух видов восстановительных операций на элементах A.

function s = multireduce(A)
  s = gpucoder.reduce(A, {@mysum, @mymax}); 
end

function c = mysum(a, b)
  c = a+b;
end

function c = mymax(a, b)
  c = max(a,b);
end

Используйте codegen для создания функции CUDA MEX.

codegen -config coder.gpuConfig('mex') -args {rand(1,1024,'double')} -report multireduce

Ниже приведен фрагмент сгенерированного кода. 

...
cudaMalloc(&gpu_s, 16ULL);
cudaMalloc(&gpu_A, 8192ULL);
cudaMemcpy(gpu_A, (void *)&A[0], 8192ULL, cudaMemcpyHostToDevice);
multireduce_kernel1<<<dim3(1U, 1U, 1U), dim3(32U, 1U, 1U)>>>(*gpu_A, *gpu_s);
coder_reduce0<<<dim3(2U, 1U, 1U), dim3(512U, 1U, 1U)>>>(*gpu_A, *gpu_s);
cudaMemcpy(&s[0], gpu_s, 16ULL, cudaMemcpyDeviceToHost);
...
static __inline__ __device__ real_T shflDown2(real_T in1, uint32_T offset,
  uint32_T mask)
{
  int2 tmp;
  tmp = *(int2 *)&in1;
  tmp.x = __shfl_down_sync(mask, tmp.x, offset);
  tmp.y = __shfl_down_sync(mask, tmp.y, offset);
  return *(real_T *)&tmp;
}
...

Входные аргументы

свернуть все

Входной массив для выполнения операции уменьшения. Для создания кода входной массив должен иметь числовой или логический тип данных.

Дескриптор для определяемой пользователем функции. FUN также может быть массивом ячеек дескрипторов функций. Дескриптор функции является двоичной функцией и должен удовлетворять следующим требованиям:

  • Примите два входа и вернете один выход. Тип входов и выходов функции должен соответствовать типу входного массива A.

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

Выходные аргументы

свернуть все

Результат операции сокращения. Во время сокращения, S инициализируется до значения одного из элементов входного массива A. Затем выполняется операция сокращения путем применения FUN каждому элементу в A и S.

Ограничения

  • gpucoder.reduce не поддерживает входные массивы сложного типа данных.

  • Определяемая пользователем функция должна принимать два входа и возвращать один выход. Тип входов и выходов функции должен соответствовать типу входного массива A.

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

  • Для некоторых входных данных, относящихся к целочисленному типу данных, сгенерированный код может содержать промежуточные вычисления, достигающие насыщения. В таких случаях результаты сгенерированного кода могут не совпадать с результатами моделирования из MATLAB ®.

См. также

Приложения

Функции

Объекты

Темы

Представлен в R2019b

Документация по кодеру GPU

Поддержка