cuDSS General Description#

This section describes general aspects of using cuDSS API.

Error status#

All cuDSS calls return the error status cudssStatus_t.

CUDA error state#

This contract applies when the application and cuDSS share a CUDA runtime instance. CUDA error state in a separately linked runtime instance is not visible to cuDSS.

Every public API call checks the CUDA runtime error state visible to cuDSS. A pending error on entry is consumed and reported before a regular API performs its operation. Destroy APIs still attempt to release their resources, and distributed APIs may perform the communication needed to agree an error status. If an otherwise successful operation leaves an error pending, the error is consumed and reflected in the returned status. Distributed cudssExecute() calls and collective cudssDataGet() calls include errors observed before their final rank agreement. An error raised by the communication used for that agreement is reported by the affected rank. Collective cudssDataGet() calls preserve an agreed argument or object-validation status. A failure arising while performing the query is reported on all ranks as CUDSS_STATUS_EXECUTION_FAILED. Property, logger, configuration, and matrix descriptor APIs ignore cudaErrorNoDevice and cudaErrorInsufficientDriver. These conditions may remain observable while the CUDA runtime is unavailable. In MGMN mode, cudssExecute() reports a general rank-local failure on all ranks as CUDSS_STATUS_EXECUTION_FAILED. CUDSS_STATUS_NOT_SUPPORTED is preserved only when no rank reports a general failure. CUDSS_STATUS_IR_FAILED is preserved on all ranks when no general or not-supported failure occurs.

CUDA memory allocation errors map to CUDSS_STATUS_ALLOC_FAILED; other CUDA errors map to CUDSS_STATUS_EXECUTION_FAILED. An operation’s non-success status takes precedence over a CUDA error observed on exit.

Checking CUDA error state may initialize CUDA Runtime state, including for APIs that do not require a GPU.

cuDSS does not synchronize solely to inspect error state, so asynchronous errors can surface in a later call and may originate outside cuDSS. Consuming an error does not recover a CUDA context from a fatal error.

C++ exceptions do not cross the public C API. std::bad_alloc maps to CUDSS_STATUS_ALLOC_FAILED; other exceptions map to CUDSS_STATUS_INTERNAL_ERROR.

User callbacks must not throw. As defensive containment, cuDSS converts an exception that reaches a public API boundary, but in multi-process execution an exception can prevent ranks from entering the same communication operation and can hang the operation.

Library handle#

In order to use cuDSS, the calling applications must create a cuDSS library handle by calling cudssCreate() function or cudssCreateMg() in case of multiple devices before calling any cuDSS API which takes the handle as an argument. Once the application finishes using the library, it must call the function cudssDestroy() to release the resources associated with the cuDSS library handle.

Memory ownership and lifetime#

cuDSS does not take ownership of data buffers provided by the user. This includes buffers used when creating cudssMatrix_t objects or user permutation or any other buffer passed via cudssDataSet() function. Array parameters passed via cudssDataSet() (CUDSS_DATA_USER_PERM, CUDSS_DATA_USER_ND_PARTITION_TREE, CUDSS_DATA_USER_SCHUR_INDICES and CUDSS_DATA_UBATCH_MASK) are copied to an internally managed memory buffer. After the call to cudssDataSet() returns, cuDSS does not require access to the provided buffer and always operates on the internal copy. In contrast, for CUDSS_DATA_USER_HOST_INTERRUPT, CUDSS_DATA_COMM_DEVICE and CUDSS_DATA_COMM_HOST cuDSS stores the provided pointer itself, not a copy of the object it points to. That object must remain valid until the parameter is set again or the cudssData_t object is destroyed. In general, if a buffer has been allocated outside of cuDSS, it will not be deallocated by cuDSS and it is the responsibility of the user to do it. Also, it is the responsibility of the user to ensure that the user-provided buffers, such as matrix and vector data, are not destroyed before cuDSS stops operating on them. Since some of cuDSS APIs (e.g. factorization and solve phases) are asynchronous, a stream synchronization via cudaStreamSynchronize() may be necessary to ensure the buffers are not in use anymore before destroying them.

cuDSS allocates host and device buffers for internal use. These buffers are kept inside the corresponding API objects like cudssHandle_t, or cudssData_t. These internal buffers are deallocated when the containing object (cudssHandle_t or cudssData_t) is destroyed during the call to cudssDestroy() or cudssDataDestroy().

cuDSS allocates host buffers when opaque objects are created via Create() functions. To deallocate those buffers, corresponding Destroy() functions must be called. A Create() call must always be paired with a Destroy() call to prevent memory issues. If a Create() call returns an error after writing a non-NULL object, the object must still be passed to the corresponding Destroy() function.

cudssMatrix_t should be thought of as a very thin wrapper around the user provided buffers.

Controlling the device buffers allocated by cuDSS internally can be done via cudssDeviceMemHandler_t and associated APIs. Optionally, cuDSS can use pinned (page-locked) host buffers, controlled via cudssPinnedMemHandler_t and associated APIs. See User-defined memory handlers for details.

Memory alignment#

For user-provided buffers, cuDSS does not have any additional memory alignment requirements, i.e., only the default data-type dependent alignment must be present.

Thread Safety#

Thread safety is not guaranteed for calling cuDSS from multiple host threads. For example, the cudssHandle_t does internal book-keeping for device buffers allocated by the library and cannot be operated upon by multiple host threads when device buffers are being allocated. Also, updates to cudssData_t objects which happen during analysis and factorization phases are not thread-safe.

Logging Features are not thread-safe either. Logger API calls must not be made from multiple threads, and logging environment variables should be set before the application starts rather than modified at runtime.

The only phase of cudssExecute() which allocates no device memory and can be executed concurrently with several host threads is the solve phase. However, in this case the host threads must operate on different cudssMatrix_t objects for the solution.

Results Reproducibility#

By default, cuDSS makes use of atomic operations which implies that bit-wise reproducibility is not guaranteed even in the repeated runs within the fixed environment. Usually such run-to-run variations should only lead to small variations in the conventional residual norms.

To enable bit-wise reproducibility in cuDSS at every run when executed on GPUs with the same architecture and the same number of SMs (assuming the input data and solver settings are also bit-wise identical), one can use the setting CUDSS_CONFIG_DETERMINISTIC_MODE.

Note: deterministic mode uses a different set of kernels which often might be slower than the kernels used in the default mode.

Parallelism with Streams#

Typically, for solving a linear system, cuDSS uses both host and device compute resources. Specifically, when the hybrid host/device execution mode is disabled, reordering (a major part of the analysis phase) is executed on the host, while symbolic factorization (another part of the analysis phase), numerical factorization and solve are executed on the GPU. Thus, quite naturally the analysis phase is always synchronous currently. Factorization and solve are generally asynchronous in single-GPU device execution. Hybrid host/device memory mode, hybrid host/device execution mode, MG mode, and MGMN mode are synchronous. Solves using CUDSS_PIVOT_BUNCH_KAUFMAN are also synchronous. See CUDA Graphs support for capture restrictions.

For configurations that support asynchronous execution, the application should first call the analysis phase on all streams (which will not be concurrently executed), and only then call (re-)factorization and solve phases in each of streams.

CUDA Graphs support#

There are restrictions on using CUDA Graphs with cuDSS.

By default, cuDSS uses cudaMalloc/cudaFree for device memory allocations which is not compatible with CUDA graphs. To remove this limitation, one should use cudssDeviceMemHandler and associated APIs so that cuDSS uses cudaMallocAsync/cudaFreeAsync for device memory allocations.
Analysis phase is always synchronous.
Solves using CUDSS_PIVOT_BUNCH_KAUFMAN are synchronous and cannot be stream-captured.
When certain features are enabled, cuDSS factorization and solve phases may also become synchronous. For example, this happens when any of hybrid host/device memory mode, hybrid host/device execution mode, MG mode, or MGMN mode is enabled.

Factorization and solve phases that support asynchronous execution can be captured in a CUDA Graph subject to these restrictions.

A captured graph references device buffers of the cudssData object and runtime-compiled kernels owned by the cudssHandle used during capture. It must not be launched after either object is destroyed, or after the analysis phase is run again with that handle.

Runtime kernel compilation#

Some cuDSS kernels are compiled at runtime with NVRTC for the GPU in use, instead of being shipped precompiled for every supported architecture. Compilation happens during the analysis phase, once per process and GPU architecture; the compiled kernels are loaded into the cudssHandle for each of its devices and released by cudssDestroy(). NVRTC is therefore a load-time dependency of cuDSS (see Getting Started), and a CUDA graph that captured cuDSS work references these kernels (see CUDA Graphs support).

CUDSS_JIT_CACHE_DIR names a directory for an on-disk cache of these kernels. When it is unset or empty (the default) there is no on-disk cache and kernels are compiled in memory, once per process. When it is set, cuDSS creates the directory if needed and stores one file per compiled kernel in it. The file names include the GPU architecture and the NVRTC version, so one directory can be shared by processes using different CUDA toolkits and GPUs; an entry that cannot be loaded (corrupt, or produced by an NVRTC release the installed driver does not accept) is replaced automatically, and the directory may be deleted at any time. If the directory cannot be created or written to, an error is logged once and kernels are compiled in memory as if the variable were unset.

The directory must be writable only by users who are trusted to run GPU code in the processes that use it: like any on-disk code cache, cuDSS loads the compiled kernels it finds there without further authentication. Do not point it at a shared location such as /tmp. On Linux cuDSS creates the directory with mode 0700 and the files in it with mode 0644 (before the umask), so no umask can make them writable by other users; on Windows a created directory inherits the permissions of its parent, so choose a location under the user’s profile.

Applications that must not depend on the environment can set the directory per library handle with cudssSetJitCacheDir(), which takes precedence over CUDSS_JIT_CACHE_DIR: an empty string disables the on-disk cache for that handle, a path selects the directory, and NULL restores the default of reading the environment variable. The setting applies to the analysis phases executed with that handle after the call. Compiled kernels are shared by all handles in a process, so a kernel another handle already compiled or loaded is reused whatever that handle’s setting was; the directory set for a handle is read when that handle is the first to need a kernel and receives every kernel the handle uses.

CUDSS_JIT_DISABLE=1 turns runtime compilation off: nothing is compiled or loaded and cuDSS uses its built-in code paths instead. This is independent of CUDSS_JIT_CACHE_DIR (unsetting the directory only disables the on-disk cache) and does not remove the load-time dependency on the NVRTC library. Features that exist only as runtime-compiled kernels are unavailable when it is set.

Environment variables#

The following environment variables are cuDSS-specific:

Environment variable

Supported values

Description

CUDSS_LOG_LEVEL

0 - 5

Controls logging level

CUDSS_LOG_MASK

0 - 31

Controls logging mask

CUDSS_LOG_FILE

string

Controls logging output file name

CUDSS_COMM_LIB

string

Name or path for the communication layer library

CUDSS_THREADING_LIB

string

Name or path for the threading layer library

CUDSS_JIT_CACHE_DIR

string

Directory for the on-disk kernel cache

CUDSS_JIT_DISABLE

0 - 1

Set to 1 to disable runtime kernel compilation

Environment variable	Supported values	Description
`CUDSS_LOG_LEVEL`	`0 - 5`	Controls logging level
`CUDSS_LOG_MASK`	`0 - 31`	Controls logging mask
`CUDSS_LOG_FILE`	`string`	Controls logging output file name
`CUDSS_COMM_LIB`	`string`	Name or path for the communication layer library
`CUDSS_THREADING_LIB`	`string`	Name or path for the threading layer library
`CUDSS_JIT_CACHE_DIR`	`string`	Directory for the on-disk kernel cache
`CUDSS_JIT_DISABLE`	`0 - 1`	Set to 1 to disable runtime kernel compilation

Fortran interface#

The cuDSS API is a C API, so calling it from Fortran requires interoperable interfaces built on iso_c_binding. For this purpose cuDSS ships a set of Fortran interface modules (shipped under include/fortran/) together with example programs under examples/fortran/ and examples/nvfortran/.

The interface is provided by the following modules:

cudss (include/fortran/cudss.f90): the main Fortran interface to the cuDSS C API. It exposes the cuDSS constants, derived types, enumerators, and interoperable procedure interfaces for creating handles, configuration and data objects, and matrix descriptors, and for calling the analysis, factorization, and solve phases.
cudss_mpi (include/fortran/cudss_mpi.f90 with include/fortran/cudss_fortran_mpi.c): a small support module for the MGMN mode that passes a Fortran MPI communicator to cuDSS.
cudautils (cudautils.f90): a minimal interface to the CUDA runtime calls used by the examples (allocation, memory copy, streams, and device selection). It is only needed when building with a non-CUDA-Fortran compiler; with nvfortran the cudafor module is used instead.

Two sets of examples are provided:

examples/fortran/ contains standard Fortran examples (.f90) that can be compiled with any Fortran compiler supporting the required Fortran and C interoperability features.
examples/nvfortran/ contains CUDA Fortran examples (.cuf) that use CUDA Fortran language features and require nvfortran.

A cuDSS Fortran program follows the same flow as the C API: create a handle with cudssCreate(), create the configuration, data, and matrix objects, and call cudssExecute() for the analysis, factorization, and solve phases.

Because a Fortran .mod file is compiler- and version-specific, the interface is shipped as source under include/fortran/ and you compile it into your own build. A minimal, non-MPI program is compiled and linked as follows (FC is your Fortran compiler, e.g. nvfortran, gfortran, or ifx; CUDSS_INC, CUDSS_LIB, and CUDA_LIB point at the cuDSS headers, the cuDSS library, and the CUDA runtime, cuBLAS, NVRTC, and NVRTC builtins libraries):

export LD_LIBRARY_PATH="$CUDSS_LIB:$CUDA_LIB${LD_LIBRARY_PATH:+:$LD_LIBRARY_PATH}"
$FC -c $CUDSS_INC/fortran/cudss.f90                  # -> cudss.mod + cudss.o
$FC -I. -c my_app.f90                        # my_app uses `use cudss`
$FC my_app.o cudss.o -L$CUDSS_LIB -lcudss \
    -L$CUDA_LIB -lcudart -lcublas -Wl,-rpath,$CUDSS_LIB -o my_app

For the MGMN mode, additionally compile the MPI support module and its C thunk (the thunk must be built with your MPI C compiler because it converts the Fortran communicator via MPI_Comm_f2c, and it includes cudss.h so its compile must see the cuDSS and CUDA headers, with CUDA_INC pointing at the CUDA headers), and build and link with the MPI Fortran compiler wrapper:

export LD_LIBRARY_PATH="$CUDSS_LIB:$CUDA_LIB${LD_LIBRARY_PATH:+:$LD_LIBRARY_PATH}"
mpicc  -I$CUDSS_INC -I$CUDA_INC -c $CUDSS_INC/fortran/cudss_fortran_mpi.c  # -> cudss_fortran_mpi.o
mpif90 -c $CUDSS_INC/fortran/cudss_mpi.f90           # -> cudss_mpi.mod + cudss_mpi.o
mpif90 my_mpi_app.o cudss.o cudss_mpi.o cudss_fortran_mpi.o \
       -L$CUDSS_LIB -lcudss -L$CUDA_LIB -lcudart -lcublas \
       -Wl,-rpath,$CUDSS_LIB -o my_mpi_app

These examples are built with the provided Makefiles (Makefile for the standard Fortran examples and the examples/nvfortran/ Makefile for the CUDA Fortran examples) and are not part of the CMake examples build. See the examples/fortran/README.md and examples/nvfortran/README.md files for the full module descriptions and for build and run instructions.

Relation to cuSolverSp and cuSolverRf components of cuSolver library#

Sparse direct solver functionality provided by cuDSS is also partially available in the (deprecated) routines from cuSolverSp and cuSolverRf components of cuSolver library. Users of cuSolverSp and cuSolverRf are strongly recommended to switch to cuDSS. For technical details, please refer to documentation for cuSolverSp and cuSolverRf and for the transition samples.