GPU Coder™ предоставляет вам доступ к двум различным выделениям памяти (malloc
) режимы, доступные в модели программирования CUDA®, cudaMalloc
и cudaMallocManaged
. cudaMalloc
API применим к традиционно отдельному центральному процессору и глобальным памятям графического процессора. cudaMallocManaged
применимо к Unified Memory.
С точки зрения программиста традиционная архитектура ЭВМ требует, чтобы данные были выделены и совместно использованы пространствами памяти центрального процессора и графического процессора. Потребность в приложениях, чтобы справиться с передачами данных между этими двумя пространствами памяти добавляет к увеличенной сложности. Объединенная память создает пул управляемой памяти, совместно использованной центральным процессором и графическим процессором. Управляемая память доступна и для центрального процессора и для графического процессора через один указатель. Объединенная память пытается оптимизировать производительность памяти путем миграции данных на устройство, которому нужен он, одновременно скрывая детали миграции от программы. Хотя объединенная память упрощает модель программирования, она требует синхронизирующих устройством вызовов, когда к данным, записанным на графическом процессоре, получают доступ на центральном процессоре. GPU Coder вставляет эти вызовы синхронизации. Согласно NVIDIA®, объединенная память может обеспечить значительные выигрыши в производительности, когда при помощи CUDA 8.0, или когда предназначение встроило оборудование как NVIDIA Tegra®.
Чтобы изменить режим выделения памяти в приложении GPU Coder, используйте Malloc Mode
выпадающее поле под More Settings->GPU Coder. При использовании интерфейса командной строки используйте MallocMode
свойство конфигурации сборки и набор это к любому 'discrete'
или 'unified'
.
GPU Coder анализирует зависимость по данным между разделами центрального процессора и графического процессора и выполняет оптимизацию, чтобы минимизировать количество cudaMemcpy
вызовы функции в сгенерированном коде. Анализ также определяет минимальный набор местоположений, где данные должны быть скопированы между центральным процессором и графическим процессором при помощи cudaMemcpy
.
Например, функциональный foo
имеет разделы кода, которые обрабатывают данные последовательно на центральном процессоре и параллельно на графическом процессоре.
function [out] = foo(input1,input2) … % CPU work input1 = … input2 = … tmp1 = … tmp2 = … … % GPU work kernel1(gpuInput1, gpuTmp1); kernel2(gpuInput2, gpuTmp1, gpuTmp2); kernel3(gpuTmp1, gpuTmp2, gpuOut); … % CPU work … = out end
Неоптимизированная реализация CUDA может потенциально иметь несколько cudaMemcpy
вызовы функции передать все входные параметры gpuInput1,gpuInput2
, и временные результаты gpuTmp1,gpuTmp2
между вызовами ядра. Поскольку промежуточное звено заканчивается gpuTmp1,gpuTmp2
не используются вне графического процессора, они могут храниться в памяти графического процессора, приводящей к меньшему количеству cudaMemcpy
вызовы функции. Эта оптимизация улучшает общую производительность сгенерированного кода. Оптимизированная реализация:
gpuInput1 = input1; gpuInput2 = input2; kernel1<<< >>>(gpuInput1, gpuTmp1); kernel2<<< >>>(gpuInput2, gpuTmp1, gpuTmp2); kernel3<<< >>>(gpuTmp1, gpuTmp2, gpuOut); out = gpuOut;
Устранить избыточный cudaMemcpy
вызовы, GPU Coder анализирует все использование и определения данной переменной и использует флаги состояния, чтобы выполнить минимизацию. Пример оригинального кода и на что похож сгенерированный код, показывают в этой таблице.
Оригинальный код | Оптимизированный сгенерированный код |
---|---|
A(:) = … … for i = 1:N gB = kernel1(gA); gA = kernel2(gB); if (somecondition) gC = kernel3(gA, gB); end … end … … = C; |
A(:) = … A_isDirtyOnCpu = true; … for i = 1:N if (A_isDirtyOnCpu) gA = A; A_isDirtyOnCpu = false; end gB = kernel1(gA); gA = kernel2(gB); if (somecondition) gC = kernel3(gA, gB); C_isDirtyOnGpu = true; end … end … if (C_isDirtyOnGpu) C = gC; C_isDirtyOnGpu = false; end … = C; |
_isDirtyOnCpu
флаг говорит оптимизацию памяти GPU Coder о стандартных программах, где данная переменная объявляется и использовала или на центральном процессоре или на затем графическом процессоре.