coder.gpu.persistentMemory - Pragma to allocate a variable as persistent memory on the GPU - MATLAB (original) (raw)

Pragma to allocate a variable as persistent memory on the GPU

Since R2020b

Syntax

Description

coder.gpu.persistentMemory([pvar](#mw%5F77aaf190-9d31-4a6f-86e8-7cf34656329c)) maps the persistent MATLAB® variable pvar to the CUDA® enabled NVIDIA® GPU as variable with persistent memory. The variable must be fixed size and of a data type supported for GPU code generation.

This is a code generation function. It has no effect in MATLAB.

Note

For CUDA MEX, the persistent memory on the GPU is retained during the entirety of the MATLAB session. For freeing the GPU memory, use the clear mex MATLAB command. For freeing the GPU memory in static libraries, dynamic libraries, or executable targets, call the generated <primary function name>_terminate() housekeeping function.

example

Examples

collapse all

This example shows how to map a persistent MATLAB variable to the persistent memory on the GPU by using thecoder.gpu.persistentMemory pragma.

Consider the following MATLAB entry-point function myPersistent containing a simple loop.

function output = myPersistent(input)

persistent pvar;
if isempty(pvar)
    pvar = ones(size(input));
end

coder.gpu.persistentMemory(pvar);

for i = 1:numel(input)
    pvar(i) = pvar(i) + input(i);
end

output = coder.nullcopy(input);
for i = 1:numel(input)
    output(i) = pvar(i) * input(i);
end

end

Create a code generation configuration object for a standalone CUDA static library.

cfg = coder.gpuConfig('lib');

Define a cell array input that declares the size and data type of the inputs to the function myPersistent.

Generate a MEX function myPersistent_mex by using codegen function with the -config,-args, and -report options to specify configuration, provide input arguments, and generate a code generation report.

codegen -config cfg -args input -report myPersistent

A snippet of the generated myPersistent.cu file is shown.

// // File: myPersistent.cu // // GPU Coder version : 2.0 // CUDA/C/C++ source code generated on : 16-Jul-2020 20:08:46 //

// Include Files #include "myPersistent.h" #include "myPersistent_data.h" #include "myPersistent_initialize.h" #include "MWCudaDimUtility.hpp" ... // // Arguments : const double input[1024] // double output[1024] // Return Type : void // void myPersistent(const double input[1024], double output[1024]) { double (*gpu_input)[1024]; double (*gpu_output)[1024]; if (!isInitialized_myPersistent) { myPersistent_initialize(); }

cudaMalloc(&gpu_output, 8192UL); cudaMalloc(&gpu_input, 8192UL); cudaMemcpy(gpu_input, (void *)&input[0], 8192UL, cudaMemcpyHostToDevice); myPersistent_kernel1<<<dim3(2U, 1U, 1U), dim3(512U, 1U, 1U)>>>(*gpu_input, *gpu_output, *pvar); cudaMemcpy(&output[0], gpu_output, 8192UL, cudaMemcpyDeviceToHost); cudaFree(*gpu_input); cudaFree(*gpu_output); }

// // Arguments : void // Return Type : void // void myPersistent_init() { double b_pvar[1024]; boolean_T pvar_dirtyOnCpu; boolean_T pvar_dirtyOnGpu; pvar_dirtyOnCpu = false; pvar_dirtyOnGpu = true; for (int i = 0; i < 1024; i++) { if (pvar_dirtyOnGpu) { cudaMemcpy(&b_pvar[0], pvar, 8192UL, cudaMemcpyDeviceToHost); pvar_dirtyOnGpu = false; }

b_pvar[i] = 1.0;
pvar_dirtyOnCpu = true;

}

if (pvar_dirtyOnCpu) { cudaMemcpy(pvar, &b_pvar[0], 8192UL, cudaMemcpyHostToDevice); } } ...

The persistent variable pvar is retained on the GPU between calls to the myPersistent() function.

Input Arguments

collapse all

The name of the variable that must be mapped to the GPU memory space as a persistent variable.

Limitations

Version History

Introduced in R2020b

expand all

The coder.gpu.persistentMemory pragma is supported for Simulink®. You can observe improved performance in Simulink simulations and generate optimized code for Simulink models.

See Also

Apps

Functions

Objects

Topics