NVIDIA NPPDx Documentation#

NVIDIA Pipeline Processing Device Extensions (NPPDx) is a header-only CUDA C++ library for fused image-processing kernels. It uses the same operator programming model as other MathDx libraries: compose operators with + at compile time, then call execute() from a user __global__ function.

Goal: Minimize global memory traffic and keep data as close to the chip as possible.

Fusion#

A fused kernel has a minimum of two processing steps: ingest and exgest for format conversion.

Image processing step categories that can be fused:

  • Ingest and Exgest: global memory access using any number of memory formats \(\rightarrow\) internal 32-bit float data on chip.

  • Pointwise process: color space changes, affine mappings, gamma transforms, arithmetic, etc.

  • Area process: box blur, Gaussian blur, sharpen, median, etc.

  • Resize: nearest, bilinear, bicubic, and Lanczos interpolation (technically an area process, but has a different output size than input size).

NPPDx is able to fuse all of these categories of processing steps, but any fused kernel that uses an area process requires halo management. At compile time the halo requirements for each step are available as traits (see also Halos and step order).

Fused kernels that do not require halo management have the option of being built with the faster Register-only API.

NPPDx is not limited to a single ingest or exgest step.

Shared-memory fused kernel showing ingest, median, Laplacian, color conversion, and exgest with shrinking tile sizes

Figure 1 Example Shared-memory fused kernel: ingest RGB16, median 5x5, additive Laplacian 3x3, color conversion (BT.601), and exgest UYVP (10-bit 4:2:2). Neighborhood processing steps consume halo so the valid interior shrinks from 7x7 to 3x3 to 1x1.#

NPPDx Library Highlights#

  • Header-only API: CUDA C++ with C++17 support.

  • Composable operators: MathDx-style + assembly for image-processing fused kernels.

  • Ingest/exgest: reading image data into registers or shared memory and writing processed data back to image buffers.

  • Tile processing: Register-only API and Shared-memory API.

  • Texture/surfaces buffer: hardware acceleration ingest and exgest for CUDA array-based workflows.

  • Color conversion: between RGB and YUV color spaces.

  • Pointwise process: gamma transforms and affine channel mapping.

  • Area process: box blur, Gaussian blur, sharpen, and median.

  • Resize: methods available are nearest neighbor, bilinear, bicubic, and Lanczos3 interpolation.

  • CMake integration: through find_package(nppdx CONFIG).

The documentation also includes the sections Requirements and Functionality, Installation Guide, Using NPPDx, Achieving High Performance, Operators, Processing overview, Traits, Execution Methods, and Examples.

The NPPDx GitHub repository is available at NVIDIA/NPPDx.