Using GPUs

GPUs (Graphics Processing Units) are chips that provide specialized parallel processing power. Originally, GPUs were designed to handle graphics processing as part of the video processor, but their ability to handle non-graphics tasks in a similar manner has become important for general computing. GPUs designed for general purpose computing task are commonly called General Purpose GPUs, or GPGPUs.

A GPU is suited for processing an algorithm that naturally breaks down into a process requiring many similar calculations running in parallel. GPUs cores are able to rapidly apply the instruction on multiple data points organized in a 2-D, and more recently, 3-D, image. The image is placed in a framebuffer. In the original chips, the data points held in the framebuffer were intended for output to a display, thereby accelerating image generation.

The similarity between multicore CPU chips and modern GPUs makes it at first sight attractive to use GPUs for general purpose computing. However, the instruction set on GPGPUs is used in a component called the shader pipeline. This has, as the name suggests, to do with a limited set of graphics operations, and so is by its nature rather limited. Using the instruction set for problems unrelated to shader pipeline manipulations requires that the problems being processed map over to a similar manipulation. This works best for algorithms that naturally break down into a process requiring an operation to be applied in the same way on many independent vertices and pixels. In practice, this means that 1-D vector operations are an order of magnitude less efficient on GPUs than operations on triangular matrices.

Modern GPGPU implementations have matured so that they can now sub-divide their resources between independent processes that work on independent data, and they provide programmer-friendlier ways of data transfer between the host and GPU memory.

Physically, one GPU is typically a built-in part of the motherboard of a node or a board in a node, and consists of several hundred processing cores. There are also dedicated standalone units, commonly called GPU Units, consisting of several GPUs in one chassis. Several of these can be assigned to particular nodes, typically via PCI-Express connections, to increase the density of parallelism even further.

NVIDIA Base Command Manager has several tools that can be used to set up and program GPUs for general purpose computations.

6.1 Packages

A number of different GPU-related packages are included in BCM. The CUDA versions supported are listed in section 9.1 of the Installation Manual.

The version implemented depends on how the system administrator has installed and configured CUDA.

6.2 Using CUDA

After installation of the packages, for general usage and compilation it is sufficient to load just the CUDA<version>/toolkit module, where <version> indicates the CUDA version number. At the time of writing of this section (May 2023), for Rocky Linux version 9.1, the versions available are 11.7, 11.8, 12.0, and 12.1.

The toolkit comes with the necessary tools and the NVIDIA compiler wrapper to compile CUDA C code.

Extensive documentation on how to get started, the various tools, and how to use the CUDA suite is at https://docs.nvidia.com/cuda.

Also available are several other modules related to CUDA:

• cuda12.1/blas: Provides paths and settings for the CUBLAS library.

• cuda12.1/fft: Provides paths and settings for the CUFFT library.

6.3 Using OpenCL

OpenCL functionality is provided with the cuda <version>/toolkit environment module.

Examples of OpenCL code can be found in the $CUDA_SDK/OpenCL directory.

6.4 Compiling Code

Both CUDA and OpenCL involve running code on different platforms:

• host: with one or more CPUs

• device: with one or more CUDA enabled GPUs

Accordingly, both the host and device manage their own memory space, and it is possible to copy data between them. The CUDA and OpenCL Best Practices Guides in the doc directory, provided by the CUDA toolkit package, have more information on how to handle both platforms and their limitations.

The nvcc command by default compiles code and links the objects for both the host system and the GPU. The nvcc command distinguishes between the two and it can hide the details from the developer. To compile the host code, nvcc will use gcc automatically.

nvcc [options] <inputfile>

A simple example to compile CUDA code to an executable is:

nvcc testcode.cu -o testcode

The most used options are:

• -g or --debug <level>: This generates debuggable code for the host

• -G or --device-debug <level>: This generates debuggable code for the GPU

• -o or --output-file <file>: This creates an executable with the name <file>

• -arch=sm_13: This can be enabled if the CUDA device supports compute capability 1.3, which includes double-precision

If double-precision floating-point is not supported or the flag is not set, warnings such as the following will come up:

warning : Double is not supported. Demoting to float

The nvcc documentation manual, “The CUDA Compiler Driver NVCC” has more information on compiler options.

The CUDA SDK has more programming examples and information accessible from the file $CUDA_SDK/C/Samples.html.

For OpenCL, code compilation can be done by linking against the OpenCL library:

  gcc test.c -lOpenCL
  g++ test.cpp -lOpenCL
  nvcc test.c -lOpenCL

6.5 Available Tools

6.5.1 CUDA gdb

The CUDA debugger can be started using: cuda-gdb. Details of how to use it are available in the “CUDA-GDB (NVIDIA CUDA Debugger)” manual, in the doc directory. It is based on GDB, the GNU Project debugger, and requires the use of the “-g” or “-G” options compiling.

Example

nvcc -g -G testcode.cu -o testcode

6.5.2 The nvidia-smi Utility

The NVIDIA System Management Interface command, nvidia-smi, can be used to allow exclusive access to the GPU. This means only one application can run on a GPU. By default, a GPU will allow multiple running applications.

Syntax:

    nvidia-smi [OPTION1 [ARG1]] [OPTION2 [ARG2]] ...

The steps for making a GPU exclusive:

• List GPUs

• Select a GPU

• Lock GPU to a compute mode

• After use, release the GPU

After setting the compute rule on the GPU, the first application that executes on the GPU blocks out any of the others attempting to run. The first application does not have to be from the user that set the exclusivity lock on the GPU.

To list the GPUs, the -L argument can be used:

  $ nvidia-smi -L
  GPU 0: (05E710DE:068F10DE)  Tesla T10 Processor  (S/N: 706539258209)
  GPU 1: (05E710DE:068F10DE)  Tesla T10 Processor  (S/N: 2486719292433)

To set the ruleset on the GPU, the -c|--compute-mode option can be used:

  $ nvidia-smi -i 0 -c 1

The ruleset may be one of the following:

• 0 - Default mode (multiple applications allowed on the GPU)

• 1 - Exclusive thread mode (only one compute context is allowed to run on the GPU, usable from one thread at a time)

• 2 - Prohibited mode (no compute contexts are allowed to run on the GPU)

• 3 - Exclusive process mode (only one compute context is allowed to run on the GPU, usable from multiple threads at a time)

To check the state of the GPU, the -q|--query option can be used:

  $ nvidia-smi -i 0 -q
  COMPUTE mode rules for GPU 0: 1

In this example, GPU0 is locked, and there is a running application using GPU0. A second application attempting to run on this GPU will not be able to run on this GPU.

  $ histogram  --device=0
  main.cpp(101) : cudaSafeCall() Runtime API error :
  no CUDA-capable device is available.

After use, the GPU can be unlocked to allow multiple users:

  nvidia-smi -i 0 -c 0

6.5.3 CUDA Utility Library

CUTIL is a simple utility library designed for use in the CUDA SDK samples. There are 2 parts for CUDA and OpenCL. The locations are:

• $$CUDA_SDK/C/lib

• $$CUDA_SDK/OpenCL/common/lib

Other applications may also refer to them, and the toolkit libraries have already been pre-configured accordingly. However, they need to be compiled prior to use. Depending on the cluster, this might have already have been done.

[fred@demo ~] cd
[fred@demo ~] cp -r $CUDA_SDK
[fred@demo ~] cd $(basename $CUDA_SDK); cd C
[fred@demo C] make
[fred@demo C] cd $(basename $CUDA_SDK); cd OpenCL
[fred@demo OpenCL] make

CUTIL provides functions for:

• parsing command line arguments

• read and writing binary files and PPM format images

• comparing data arrays (typically used for comparing GPU results with CPU results)

• timers

• macros for checking error codes

• checking for shared memory bank conflicts

6.5.4 CUDA “Hello world” Example

A hello world example code using CUDA is:

Example

/*
  CUDA example
  "Hello World" using shift13, a rot13-like function.
  Encoded on CPU, decoded on GPU.
  rot13 cycles between 26 normal alphabet characters.
  shift13 shifts 13 steps along the normal alphabet characters
  So it translates half the alphabet into non-alphabet characters
  shift13 is used because it is simpler than rot13 in c
  so we can focus on the point
  (c) NVIDIA
  Taras Shapovalov <tshapovalov@nvidia.com>
 */
#include <cuda.h>
#include <stdio.h>
// CUDA kernel definition: undo shift13
__global__ void helloWorld(char* str) {
  int idx = blockIdx.x * blockDim.x + threadIdx.x;
  str[idx] -= 13;
}
void checkCudaError(cudaError_t err, const char *msg) {
  if (err != cudaSuccess) {
    fprintf(stderr, "CUDA error: %s: %s\n", msg, cudaGetErrorString(err));
    exit(EXIT_FAILURE);
  }
}
int main(int argc, char **argv) {
  char s[] = "Hello World!";
  printf("String for encode/decode: %s\n", s);
  // CPU shift13
  int len = sizeof(s);
  for (int i = 0; i < len; i++) {
    s[i] += 13;
  }
  printf("String encoded on CPU as: %s\n", s);
  // Allocate memory on the CUDA device
  char *cuda_s;
  checkCudaError(cudaMalloc((void**)&cuda_s, len), "cudaMalloc");
  // Copy the string to the CUDA device
  checkCudaError(cudaMemcpy(cuda_s, s, len, cudaMemcpyHostToDevice), "cudaMemcpy");
  // Set the grid and block sizes (dim3 is a type)
  // and "Hello World!" is 12 characters, say 3x4
  dim3 dimGrid(3);
  dim3 dimBlock(4);
  // Invoke the kernel to undo shift13 in GPU
  helloWorld<<< dimGrid, dimBlock >>>(cuda_s);
  checkCudaError(cudaGetLastError(), "kernel execution");
  // Retrieve the results from the CUDA device
  checkCudaError(cudaMemcpy(s, cuda_s, len, cudaMemcpyDeviceToHost), "cudaMemcpy");
  // Free up the allocated memory on the CUDA device
  checkCudaError(cudaFree(cuda_s), "cudaFree");
  printf("String decoded on GPU as: %s\n", s);
  return 0;
}

The hello world code example may be compiled and run on a node with a GPU and CUDA 12.1 with:

[fred@node001 ~]$ module load shared
[fred@node001 ~]$ module load cuda12.1
[fred@node001 ~]$ nvcc hello.cu -o hellocuda
[fred@node001 ~]$ ./hellocuda
String for encode/decode: Hello World!
String encoded on CPU as: Uryy|-d|yq.
String decoded on GPU as: Hello World!
[fred@node001 ~]$

The number of characters displayed in the encoded string are less than the number in the unencoded string. This is because there are unprintable characters generated by the encoding, because the cipher used is not exactly rot13.

To make it run from a head node via a workload manager such as Slurm, on a compute node with a GPU, the following batch file could be built and run:

[fred@basecm11 ~]$ cat helloslurm.sh
#!/bin/sh
#SBATCH -o my.stdout
#SBATCH --ntasks-per-node=1
#SBATCH -p defq    #assuming node in defq has a GPU
#SBATCH --gpus=1
module clear -f
./hellocuda
[fred@basecm11 ~]$ sbatch helloslurm.sh
[fred@basecm11 ~]$ cat my.stdout
String for encode/decode: Hello World!
String encoded on CPU as: Uryy|-d|yq.
String decoded on GPU as: Hello World!

6.5.5 OpenACC

OpenACC (http://www.openacc-standard.org) is a new open parallel programming standard aiming at simplifying the programmability of heterogeneous CPU/GPU computing systems. OpenACC allows parallel programmers to provide OpenACC directives to the compiler, identifying which areas of code to accelerate. This frees the programmer from carrying out time-consuming modifications to the original code itself. By pointing out parallelism to the compiler, directives get the compiler to carry out the details of mapping the computation onto the accelerator.

Using OpenACC directives requires a compiler that supports the OpenACC standard.

In the following example, where π is calculated, adding the #pragma directive is sufficient for the compiler to produce code for the loop that can run on either the GPU or CPU:

Example

 #include <stdio.h>
  #define N 1000000
  int main(void) {
  double pi = 0.0f; long i;
  #pragma acc parallel loop reduction(+:pi)
  for (i=0; i<N; i++) {
     double t= (double)((i+0.5)/N);
     pi +=4.0/(1.0+t*t);
  }
  printf("pi=%16.15f\n",pi/N);
  return 0;
  }