To develop code for use on the NVIDIA GPU nodes, you can choose one of the three NVIDIA-recommended programming models or combine them to take advantage of the underlying NVIDIA acceleration libraries:

• Standard-language-based parallel programming model.
• Directive-based OpenACC model.
• Language-extension-based CUDA model.

Each model is briefly described in this article. Read the relevant references for each model listed at the end of this article to learn more information and limitations.

Note: NVIDIA provides limited support for using OpenMP on the GPUs. This model is not described in this article.

Compiler and CUDA Support

Compiler support for these models is provided through the NVIDIA HPC Software Development Kit (SDK), which includes these compilers:

• NVIDIA HPC Compilers for NVIDIA GPUs and Intel, AMD, OpenPower, and Arm CPUs:
  • NVFORTRAN: Fortran compiler with support for CUDA Fortran, originally developed by PGI
  • NVC: C11 compiler
  • NVC++: C++17 compiler
• NVIDIA CUDA C and C++ Compiler Driver for NVIDIA GPUs: NVCC

Note: The PGI compilers and tools were rebranded to become parts of the NVIDIA HPC SDK.

As these models are under active development or improvement, we recommend using the latest version of the NVIDIA HPC SDK, as it should have better support for new features when they are added to these programming models. Be aware that the available versions of NVIDIA HPC SDK at NAS vary between the GPU nodes with x86_64-based CPU (sky_gpu, cas_gpu, rom_gpu, mil_a100) and those with the aarch64-based CPU (gh200). Run the command module avail nvhpc to see what are available on each system before using the command module load to load one of them.

For more information on each compiler, read its man page or help page after loading a module. For example:

• man nvfortran
• nvfortran -h
• nvfortran -gpu -h

During runtime, your CUDA application requires one CUDA-capable GPU (for example, a V100, A100, or H100 GPU) and a CUDA driver that is compatible with the version of CUDA Toolkit used for building the application. Be aware that the CUDA driver is backward compatible. Therefore, an application compiled with an older CUDA Toolkit version can run with a newer CUDA driver. See CUDA Toolkit and CUDA Driver to learn more.

Note: The CUDA driver is installed by the system administrator on the GPU. Run either the nvidia-smi or the nvaccelinfo command to find the CUDA driver version installed on the system.

The Standard-Language-Based Parallel Programming Model

In this model, parallelism for running on the NVIDIA GPUs is achieved solely using features provided by the standard C++ and Fortran languages without the use of compiler directives such as OpenMP and OpenACC or language extensions such as CUDA. In addition, applications written with standard language parallelism can run on any platform—such as different multi-core CPUs or GPUs from different vendors—that supports these languages.

International Organization for Standardization (ISO) C++

The C++17 Standard introduced higher-level parallelism features that allow you to request a specific execution policy as the first parameter to a Standard Library algorithm. Most Standard C++ algorithms support the execution policies. The four available policies are:

• std::execution::seq: Sequential execution. No parallelism is allowed.
• std::execution::unseq: Vectorized execution on the calling thread (this execution policy was added in C++20).
• std::execution::par: Parallel execution on one or more threads.
• std::execution::par_unseq: Parallel execution on one or more threads, with each thread possibly vectorized.

Here is an example from NVIDIA documentation—a sort algorithm with the par execution policy:

std:sort(std::execution::par,
employees.begin(), employees.end(),
CompareByLastName());

NVC++ can compile Standard C++ algorithms with the par or the par_unseq execution policies through the use of the -stdpar compiler flag and automatically offloads C++ parallel algorithms to NVIDIA Volta, Turing, Amphere, or Hopper GPUs.

ISO Fortran

ISO Fortran 2018 introduced the DO CONCURRENT construct to allow you to express loop-level parallelism:

DO CONCURRENT (i=1:n)
  loop body
END DO

When the NVFORTRAN -stdpar option is specified, the compiler parallelizes the DO CONCURRENT loops and offloads them to the GPU.

Example codes that use this model are available in the NVIDIA HPC SDK installed on NAS systems.

For GPU nodes with x86_64-based CPU:

/nasa/nvidia/hpc_sdk/..../examples/stdpar

For GPU nodes with aarch64-based CPU:

/nasa/arm/software/nvhpc/..../examples/stdpar

You can compile code either on a GPU node; on a Cabeus front end (CFE), to run on the GPUs with x86_64-based CPUs (sky_gpu, cas_gpu, rom_gpu, mil_a100); or on cghfe01 and cghfe02 to run on the GPUs with aarch64 based CPU (gh200):

nvc++ -stdpar[=gpu] [-gpu=ccXY] program.cpp
nvfortran -stdpar[=gpu] [-gpu=ccXY] program.f90

In the event you want to compile the same code for running on multicore CPUs, use:

nvc++ -stdpar=multicore program.cpp
nvfortran -stdpar=multicore program.f90

To facilitate such standard language parallelism, NVC++ and NVFORTRAN rely on CUDA Unified Memory for all data movement between CPU and GPU memory. Through support in both the CUDA device driver and the NVIDIA GPU hardware, the CUDA Unified Memory manager automatically moves some types of data based on usage.

NVC++ and NVFORTRAN auto-detect the host CPU and device GPU architectures of the system where you perform the compilation. An executable built on one system may not run on another system if the host CPU instructions and/or the GPU instructions in the executable are not compatible between the two systems.

To generate an executable for a specific CPU architecture, add the -tp option. For example, use -tp=skylake for a CPU host with Intel Skylake or Cascade Lake processors, -tp=zen2 for the AMD Rome processor, or -tp=zen3 for the AMD Milan processor. For the GH200 GPU nodes (gh200), NVIDIA recommends using the flag ‑tp=native or ‑mcpu=native.

To generate an executable for a specific GPU architecture, add the -gpu=ccXY option, where ccXY stands for the compute capability of an architecture. For example, use -gpu=cc70 for V100 nodes, -gpu=cc80 for A100 nodes, and -gpu=cc90 for the GH200 nodes. Using multiple -gpu=ccXY options in a single compilation will generate one version of code for each architecture, but the compilation time will be longer.

Executables built for the x86_64 platform (PFE, AFE, CFE, ivy, has, bro, sky_ele, cas_ait, rom_ait, mil_ait, sky_gpu, cas_gpu, rom_gpu, and mil_100 GPU nodes) cannot be used on the aarch64 platform (gh200 GPU nodes) because x86_64 and aarch64 architectures have different instruction sets.

Note: A executable built with just -gpu=cc70 can run on an architecture with compute capability cc70 (V100) or cc80 (A100). However, an executable built with just -gpu=cc80 can run on an architecture with compute capability cc80 (A100) but not cc70 (V100).

Note: -stdpar defaults to -stdpar=gpu. If you compile the application with -stdpar on a Cabeus front end (cfe01 and cfe02, for running on the GPU nodes with x86-based CPUs) or on cghfe01 or cghfe02 (for running on the GPU nodes with aarch64-based CPUs) but if you do not explicitly specify one or more -gpu=ccXY, it will automatically enable: -gpu=cc50 -gpu=cc60 -gpu=cc61 -gpu=cc70 -gpu=cc75 -gpu=cc80 -gpu=cc86 -gpu=cc90

The Directive-Based OpenACC Model

OpenACC is a programming standard developed by PGI, CRAY, CAPS, and NVIDIA for offloading programs written in C, C++, and Fortran from a host CPU to an attached accelerator device.

Similar to OpenMP, OpenACC requires you to annotate your source code with directives describing sections of your code and corresponding data that are to be executed on the GPU:

• For C and C++, the directive is #pragma acc ...
• For Fortran, the directive is !$acc ...

Example OpenACC codes from NVIDIA are available on the NAS systems under:

For GPU nodes with x86_64-based CPUs:

/nasa/nvidia/hpc_sdk/..../examples/OpenACC

For GPU nodes with aarch64-based CPUs:

/nasa/arm/software/nvhpc/..../examples/OpenACC

To compile your OpenACC code, use the NVIDIA HPC compilers (nvfortran, nvc, or nvc++) and options on a system with NVIDIA GPUs. For example:

nvfortran -acc=gpu -fast -Minfo=all your_program.f
nvc       -acc=gpu -fast -Minfo=all your_program.c
nvc++     -acc=gpu -fast -Minfo=all your_program.cc

where -acc=gpu will parallelize and offload OpenACC regions to the NVIDIA GPUs. The option -Minfo will generate informational messages about optimization. You can specify -acc=multicore to parallelize for a multicore CPU or -acc=host to generate an executable that will run serially on the host CPU.

As in the case of the standard-language-based model described above, you can also use the -tp option to specify a target CPU architecture and/or the -gpu=ccXY option to target GPUs with the specified compute capability.

NVIDIA OpenACC supports multiple environment variables to control behavior of accelerator-enabled programs at execution. See the Environment Variables section of the OpenACC Getting Started Guide listed in the References.

Note on OpenMP Support: Although NVIDIA does not emphasize the use of OpenMP, NVIDIA HPC compilers do support a subset of the OpenMP API with a focus on OpenMP 5.0 features for running on CPUs and GPUs. Similar to compiling OpenACC code with -acc=gpu -gpu=cc80, use the compiler flags -mp=gpu -gpu=cc80 to compile OpenMP code for offloading to run on the A100 GPUs. For codes that are well-structured for the GPUs, these should compile and execute with performance equal to or close to equivalent OpenACC. Codes that are not well-structured for GPUs may perform poorly but should execute correctly. Read the OpenMP chapter of the NVIDIA HPC Compilers User's Guide for more information.

The Language-Based CUDA Model

The Compute Unified Device Architecture (CUDA) was created by NVIDIA and implemented on the NVIDIA GPUs. With CUDA, code developers write programs in C, C++, and Fortran, and incorporate CUDA extensions and optimized libraries. This is the approach to take if you want to obtain the best possible performance on the NVIDIA platform.

A CUDA code usually includes the following sequence of operations:

1. Declare and allocate the host memory (i.e., the CPU) and device (i.e. the GPU) memory.
2. Initialize host data.
3. Copy data from host to device.
4. Run the kernels on the GPUs.
5. Copy the results from device to host.
6. Free the allocated host and device memory.

You can use either of the following programming environments:

CUDA C/C++

CUDA C/C++ is the original CUDA programming environment developed by NVIDIA as extensions to the C/C++ languages for GPUs. It allows direct programming for the GPU from a high-level language. This NVIDIA technical blog provides an easy introduction to CUDA C and C++.

If you choose to write or rewrite portions of your code in CUDA C/C++, you will need to use the NVIDIA CUDA C/C++ compiler (nvcc) to build the executable. For example:

nvcc your_program.cu

The suffix .cu indicates it is a CUDA C/C++ file.

A CUDA application binary (with one or more GPU kernels) can contain compiled GPU code in two forms: (1) binary cubin objects and (2) forward-compatible PTX assembly for each kernel. During application load time, the PTX form is compiled just-in-time (JIT) to a cubin for execution.

To build an application that can run on either V100 or A100 nodes, or on a future architecture, use the -gencode option as follows:

nvcc -gencode=arch=compute_70,code=sm_70
     -gencode=arch=compute_80,code=sm_80
     -gencode=arch=compute_80,code=compute_80
     your_program.cu

where the first and second -gencode options ending with code=sm_70 and sm_80 cause the generation of a cubin targeting V100 or A100, respectively, and the last -gencode option ending with code=compute_80 allows for the generation of PTX assembly to provide compatibility of future architectures through JIT compiling.

To build an application that can run on the GH200 nodes, use:

nvcc -gencode=arch=compute_90,code=sm_90
     -gencode=arch=compute_90,code=compute_90
     your_program.cu

CUDA Fortran

CUDA Fortran is a product of a collaboration between PGI and NVIDIA. It is based on the CUDA C/C++ runtime API, and includes a Fortran 2003 compiler and tool chain for programming NVIDIA GPUs using Fortran. Read this NVIDIA Developer Blog to get a quick introduction of CUDA Fortran.

Example CUDA Fortran codes are available in the NVIDIA HPC SDK installed on NAS systems:

For GPU nodes with x86_64-based CPUs:

/nasa/nvidia/hpc_sdk/..../examples/CUDA-Fortran

For GPU nodes with aarch64-based CPUs:

/nasa/arm/software/nvhpc/..../examples/CUDA-Fortran

To use CUDA Fortran, use the compiler nvfortran to build the executable. For example:

nvfortran your_program.cuf

Files with .cuf or .CUF extensions automatically enable CUDA Fortran compilation. If the filename of your program does not end with the .cuf or .CUF suffix, use the -cuda option. For example:

nvfortran -cuda your_program.xx

As in the case of the previous two programming models described above, you can also use the -tp option to specify a target CPU architecture and/or the -gpu=ccXY option to target GPUs with the specified compute capability.

If you have any questions or need assistance, contact the NAS Control Room at (800) 331-8737, (650) 604-4444, or support@nas.nasa.gov.

References for NVIDIA HPC Compilers

• NVIDIA HPC Compilers User's Guide

References for the Standard-Language-Based Parallel Programming Model

• Video - No More Porting: Coding for GPUs with Standard C++, Fortran and Python
• Developing Accelerated Code with Standard Language Parallelism
• C++ Parallel Algorithms
• Accelerating Standard C++ with GPUs Using stdpar
• Multi-GPU programming with standard Parallel C++, Part 1
• Multi-GPU programming with standard Parallel C++, Part 2
• Accelerating Fortran DO CONCURRENT with GPUs and the NVIDIA HPC SDK
• Using Fortran Standard Parallel Programming for GPU Acceleration
• Bringing Tensor Cores to Standard Fortran

Reference for the OpenACC Programming Model

• OpenACC Getting Started Guide

References for the the Language-Extension-Based CUDA Model

• An Even Easier Introduction to CUDA
• CUDA C++ Programming Guide
• NVIDIA Amphere GPU Architecture Compatability Guide for CUDA Applications
• An Easy Introduction to CUDA Fortran
• NVIDIA CUDA Fortran Programming Guide
• Using Tensor Cores in CUDA Fortran