The Nsight Systems CLI provides a simple interface to collect on a target without using the GUI. The collected data can then be copied to any system and analyzed later.

The CLI is distributed in the Target directory of the standard Nsight Systems download package. Users who want to install the CLI as a standalone tool can do so by copying the files within the Target directory. If you want the CLI output file (.qdstrm) to be auto-converted (to .nsys-rep) after the analysis is complete, you will need to copy the host directory as well.

If you wish to run the CLI without root (recommended mode), you will want to install in a directory where you have full access.

Note that you must run the CLI on Windows as administrator.

The Nsight Systems command lines can have one of two forms:

nsys [global_option]

or

nsys [command_switch][optional command_switch_options][application] [optional application_options]

All command line options are case sensitive. For command switch options, when short options are used, the parameters should follow the switch after a space; e.g. -s process-tree. When long options are used, the switch should be followed by an equal sign and then the parameter(s); e.g. --sample=process-tree.

For this version of Nsight Systems, if you launch a process from the command line to begin analysis, the launched process will be terminated when collection is complete, including runs with --duration set, unless the user specifies the --kill none option (details below). The exception is that if the user uses NVTX, cudaProfilerStart/Stop, or hotkeys to control the duration, the application will continue unless --kill is set.

The Nsight Systems CLI supports concurrent analysis by using sessions. Each Nsight Systems session is defined by a sequence of CLI commands that define one or more collections (e.g. when and what data is collected). A session begins with either a start, launch, or profile command. A session ends with a shutdown command, when a profile command terminates, or, if requested, when all the process tree(s) launched in the session exit. Multiple sessions can run concurrently on the same system.

The Nsight Systems command line interface can be used in two modes. You may launch your application and begin analysis with options specified to the nsys profile command. Alternatively, you can control the launch of an application and data collection using interactive CLI commands.

Version Information

nsys -v

Effect: Prints tool version information to the screen.

Run with elevated privilege

sudo nsys profile <app>

Effect: Nsight Systems CLI (and target application) will run with elevated privilege. This is necessary for some features, such as FTrace or system-wide CPU sampling. If you don't want the target application to be elevated, use `--run-as` option.

Default analysis run

nsys profile <application>
    [application-arguments]

Effect: Launch the application using the given arguments. Start collecting immediately and end collection when the application stops. Trace CUDA, OpenGL, NVTX, and OS runtime libraries APIs. Collect CPU sampling information and thread scheduling information. With Nsight Systems Embedded Platforms Edition this will only analysis the single process. With Nsight Systems Workstation Edition this will trace the process tree. Generate the report#.nsys-rep file in the default location, incrementing the report number if needed to avoid overwriting any existing output files.

Limited trace only run

nsys profile --trace=cuda,nvtx -d 20
    --sample=none --cpuctxsw=none -o my_test <application>
    [application-arguments]

Effect: Launch the application using the given arguments. Start collecting immediately and end collection after 20 seconds or when the application ends. Trace CUDA and NVTX APIs. Do not collect CPU sampling information or thread scheduling information. Profile any child processes. Generate the output file as my_test.nsys-rep in the current working directory.

Delayed start run

nsys profile -e TEST_ONLY=0 -y 20
    <application> [application-arguments]

Effect: Set environment variable TEST_ONLY=0. Launch the application using the given arguments. Start collecting after 20 seconds and end collection at application exit. Trace CUDA, OpenGL, NVTX, and OS runtime libraries APIs. Collect CPU sampling and thread schedule information. Profile any child processes. Generate the report#.nsys-rep file in the default location, incrementing if needed to avoid overwriting any existing output files.

Collect ftrace events

nsys profile --ftrace=drm/drm_vblank_event
    -d 20

Effect: Collect ftrace drm_vblank_event events for 20 seconds. Generate the report#.nsys-rep file in the current working directory. Note that ftrace event collection requires running as root. To get a list of ftrace events available from the kernel, run the following:

sudo cat /sys/kernel/debug/tracing/available_events

Run GPU metric sampling on one TU10x

nsys profile --gpu-metrics-device=0
	--gpu-metrics-set=tu10x-gfxt <application>

Effect: Launch application. Collect default options and GPU metrics for the first GPU (a TU10x), using the tu10x-gfxt metric set at the default frequency (10 kHz). Profile any child processes. Generate the report#.nsys-rep file in the default location, incrementing if needed to avoid overwriting any existing output files.

Run GPU metric sampling on all GPUs at a set frequency

nsys profile --gpu-metrics-device=all
	--gpu-metrics-frequency=20000 <application>

Effect: Launch application. Collect default options and GPU metrics for all available GPUs using the first suitable metric set for each and sampling at 20 kHz. Profile any child processes. Generate the report#.nsys-rep file in the default location, incrementing if needed to avoid overwriting any existing output files.

Collect custom ETW trace using configuration file

nsys profile --etw-provider=file.JSON

Effect: Configure custom ETW collectors using the contents of file.JSON. Collect data for 20 seconds. Generate the report#.nsys-rep file in the current working directory.

nsys profile --help

Typical case: profile a Python script that uses CUDA

nsys profile --trace=cuda,cudnn,cublas,osrt,nvtx
    --delay=60 python my_dnn_script.py

Effect: Launch a Python script and start profiling it 60 seconds after the launch, tracing CUDA, cuDNN, cuBLAS, OS runtime APIs, and NVTX as well as collecting thread schedule information.

Typical case: profile an app that uses Vulkan

nsys profile --trace=vulkan,osrt,nvtx
    --delay=60 ./myapp

Effect: Launch an app and start profiling it 60 seconds after the launch, tracing Vulkan, OS runtime APIs, and NVTX as well as collecting CPU sampling and thread schedule information.

Collect from beginning of application, end manually

nsys start --stop-on-exit=false
nsys launch --trace=cuda,nvtx --sample=none <application> [application-arguments]
nsys stop

Effect: Create interactive CLI process and set it up to begin collecting as soon as an application is launched. Launch the application, set up to allow tracing of CUDA and NVTX as well as collection of thread schedule information. Stop only when explicitly requested. Generate the report#.nsys-rep in the default location.

Note:

If you start a collection and fail to stop the collection (or if you are allowing it to stop on exit, and the application runs for too long) your system’s storage space may be filled with collected data causing significant issues for the system. Nsight Systems will collect a different amount of data/sec depending on options, but in general Nsight Systems does not support runs of more than 5 minutes duration.

Run application, begin collection manually, run until process ends

nsys launch -w true <application> [application-arguments]
nsys start

Effect: Create interactive CLI and launch an application set up for default analysis. Send application output to the terminal. No data is collected until you manually start collection at area of interest. Profile until the application ends. Generate the report#.nsys-rep in the default location.

Note:

If you launch an application and that application and any descendants exit before start is called Nsight Systems will create a fully formed .nsys-rep file containing no data.

Run application, start/stop collection using cudaProfilerStart/Stop

nsys start -c cudaProfileApi
nsys launch -w true <application> [application-arguments]

Effect: Create interactive CLI process and set it up to begin collecting as soon as a cudaProfileStart() is detected. Launch application for default analysis, sending application output to the terminal. Stop collection at next call to cudaProfilerStop, when the user calls nsys stop, or when the root process terminates. Generate the report#.nsys-rep in the default location.

Note:

If you call nsys launch before nsys start -c cudaProfilerApi and the code contains a large number of short duration cudaProfilerStart/Stop pairs, Nsight Systems may be unable to process them correctly, causing a fault. This will be corrected in a future version.

Note:

The Nsight Systems CLI does not support multiple calls to the cudaProfilerStart/Stop API at this time.

Run application, start/stop collection using NVTX

nsys start -c nvtx
nsys launch -w true -p MESSAGE@DOMAIN <application> [application-arguments]

Effect: Create interactive CLI process and set it up to begin collecting as soon as an NVTX range with given message in given domain (capture range) is opened. Launch application for default analysis, sending application output to the terminal. Stop collection when all capture ranges are closed, when the user calls nsys stop, or when the root process terminates. Generate the report#.nsys-rep in the default location.

Note:

The Nsight Systems CLI only triggers the profiling session for the first capture range.

NVTX capture range can be specified:

• Message@Domain: All ranges with given message in given domain are capture ranges. For example:

  nsys launch -w true -p profiler@service ./app

  This would make the profiling start when the first range with message "profiler" is opened in domain "service".

• Message@*: All ranges with given message in all domains are capture ranges. For example:

  nsys launch -w true -p profiler@* ./app

  This would make the profiling start when the first range with message "profiler" is opened in any domain.

• Message: All ranges with given message in default domain are capture ranges. For example:

  nsys launch -w true -p profiler ./app

  This would make the profiling start when the first range with message "profiler" is opened in the default domain.

• By default only messages, provided by NVTX registered strings are considered to avoid additional overhead. To enable non-registered strings check please launch your application with NSYS_NVTX_PROFILER_REGISTER_ONLY=0 environment:

  nsys launch -w true -p profiler@service -e NSYS_NVTX_PROFILER_REGISTER_ONLY=0 ./app

Run application, start/stop collection multiple times

The interactive CLI supports multiple sequential collections per launch.

nsys launch <application> [application-arguments]
nsys start
nsys stop
nsys start
nsys stop
nsys shutdown --kill sigkill

Effect: Create interactive CLI and launch an application set up for default analysis. Send application output to the terminal. No data is collected until the start command is executed. Collect data from start until stop requested, generate report#.qstrm in the current working directory. Collect data from second start until the second stop request, generate report#.nsys-rep (incremented by one) in the current working directory. Shutdown the interactive CLI and send sigkill to the target application's process group.

Note:

Calling nsys cancel after nsys start will cancel the collection without generating a report.

Display default statistics

nsys stats report1.nsys-rep

Effect: Export an SQLite file named report1.sqlite from report1.nsys-rep (assuming it does not already exist). Print the default reports in column format to the console.

Note: The following two command sequences should present very similar information:

nsys profile --stats=true <application>

or

nsys profile <application>

nsys stats report1.nsys-rep

Display specific data from a report

nsys stats --report gputrace report1.nsys-rep

Effect: Export an SQLite file named report1.sqlite from report1.nsys-rep (assuming it does not already exist). Print the report generated by the gputrace script to the console in column format.

Generate multiple reports, in multiple formats, output multiple places

nsys stats --report gputrace --report gpukernsum --report cudaapisum --format csv,column --output .,- report1.nsys-rep

Effect: Export an SQLite file named report1.sqlite from report1.nsys-rep (assuming it does not already exist). Generate three reports. The first, the gputrace report, will be output to the file report1_gputrace.csv in CSV format. The other two reports, gpukernsum and cudaapisum, will be output to the console as columns of data. Although three reports were given, only two formats and outputs are given. To reconcile this, both the list of formats and outputs is expanded to match the list of reports by repeating the last element.

Submit report data to a command

nsys stats --report cudaapisum --format table \ --output @“grep -E (-|Name|cudaFree” test.sqlite

Effect: Open test.sqlite and run the cudaapisum script on that file. Generate table data and feed that into the command grep -E (-|Name|cudaFree). The grep command will filter out everything but the header, formatting, and the cudaFree data, and display the results to the console.

Note: When the output name starts with @, it is defined as a command. The command is run, and the output of the report is piped to the command's stdin (standard-input). The command's stdout and stderr remain attached to the console, so any output will be displayed directly to the console.

Be aware there are some limitations in how the command string is parsed. No shell expansions (including *, ?, [], and ~) are supported. The command cannot be piped to another command, nor redirected to a file using shell syntax. The command and command arguments are split on whitespace, and no quotes (within the command syntax) are supported. For commands that require complex command line syntax, it is suggested that the command be put into a shell script file, and the script designated as the output command

The nsys stats command can be used post analysis to generate specific or personalized reports. For a default fixed set of summary statistics to be automatically generated, you can use the --stats option with the nsys profile or nsys start command to generate a fixed set of useful summary statistics.

If your run traces CUDA, these include CUDA API, Kernel, and Memory Operation statistics:

If your run traces OS runtime events or NVTX push-pop ranges:

If your run traces graphics debug markers these include DX11 debug markers, DX12 debug markers, Vulkan debug markers or KHR debug markers:

Recipes for these statistics as well as documentation on how to create your own metrics will be available in a future version of the tool.

The CLI generates a .qdstrm file. The .qdstrm file is an intermediate result file, not intended for multiple imports. It needs to be processed, either by importing it into the GUI or by using the standalone QdstrmImporter to generate an optimized .nsys-rep file. Use this .nsys-rep file when re-opening the result on the same machine, opening the result on a different machine, or sharing results with teammates.

This version of Nsight Systems will attempt to automatically convert the .qdstrm file to a .nsys-rep file with the same name after the run finishes if the required libraries are available. The ability to turn off auto-conversion will be added in a later version.

Import Into the GUI

The CLI and host GUI versions must match to import a .qdstrm file successfully. The host GUI is backward compatible only with .nsys-rep files.

Copy the .qdstrm file you are interested in viewing to a system where the Nsight Systems host GUI is installed. Launch the Nsight Systems GUI. Select File->Import... and choose the .qdstrm file you wish to open.

The import of really large, multi-gigabyte, .qdstrm files may take up all of the memory on the host computer and lock up the system. This will be fixed in a later version.

Importing Windows ETL files

For Windows targets, ETL files captured with Xperf or the log.cmd command supplied with GPUView in the Windows Performance Toolkit can be imported to create reports as if they were captured with Nsight Systems's "WDDM trace" and "Custom ETW trace" features. Simply choose the .etl file from the Import dialog to convert it to a .nsys-rep file.

Create .nsys-rep Using QdstrmImporter

The CLI and QdstrmImporter versions must match to convert a .qdstrm file into a .nsys-rep file. This .nsys-rep file can then be opened in the same version or more recent versions of the GUI.

To run QdstrmImporter on the host system, find the QdstrmImporter binary in the Host-x86_64 directory in your installation. QdstrmImporter is available for all host platforms. See options below.

To run QdstrmImporter on the target system, copy the Linux Host-x86_64 directory to the target Linux system or install Nsight Systems for Linux host directly on the target. The Windows or macOS host QdstrmImporter will not work on a Linux Target. See options below.

Profiling on Windows devices is similar to the profiling on Linux devices. Please refer to the Profiling Linux Targets from the GUI section for the detailed documentation and connection information. The major differences on the platforms are listed below:

Profiling on QNX devices is similar to the profiling on Linux devices. Please refer to the Profiling Linux Targets from the GUI section for the detailed documentation. The major differences on the platforms are listed below:

• Backtrace sampling is not supported. Instead backtraces are collected for long OS runtime libraries calls. Please refer to the OS Runtime Libraries Trace section for the detailed documentation.

• CUDA support is limited to CUDA 9.0+

• Filesystem on QNX device might be mounted read-only. In that case Nsight Systems is not able to install target-side binaries, required to run the profiling session. Please make sure that target filesystem is writable before connecting to QNX target. For example, make sure the following command works:

  echo XX > /xx && ls -l /xx

Nsight Systems has the ability to export SQLite database files from the .nsys-rep results file. From the CLI, use nsys export. From the GUI, call File->Export....

Note: The .nsys-rep report format is the only data format for Nsight Systems that should be considered forward compatible. The SQLite schema can and will change in the future.

The schema for a concrete database can be obtained with the sqlite3 tool built-in command .schema. The sqlite3 tool can be located in the Target or Host directory of your Nsight Systems installation.

Note: Currently tables are created lazily, and therefore not every table described in the documentation will be present in a particular database. This will change in a future version of the product. If you want a full schema of all possible tables, use nsys export --lazy=false during export phase.

Currently, a table is created for each data type in the exported database. Since usage patterns for exported data may vary greatly and no default use cases have been established, no indexes or extra constraints are created. Instead, refer to the SQLite Examples section for a list of common recipes. This may change in a future version of the product.

To check the version of your exported SQLite file, check the value of EXPORT_SCHEMA_VERSION in the EXPORT_META_DATA table. The schema version is a common three-value major/minor/micro version number. The first value, or major value, indicates the overall format of the database, and is only changed if there is a major re-write or re-factor of the entire database format. It is assumed that if the major version changes, all scripts or queries will break. The middle, or minor, version is changed anytime there is a more localized, but potentially breaking change, such as renaming an existing column, or changing the type of an existing column. The last, or micro version is changed any time there are additions, such as a new table or column, that should not introduce any breaking change when used with well-written, best-practices queries.

This is the schema as of the 2021.5 release, schema version 2.7.1.

CREATE TABLE StringIds (
    -- Consolidation of repetitive string values.

    id                          INTEGER   NOT NULL   PRIMARY KEY,      -- ID reference value.
    value                       TEXT      NOT NULL                     -- String value.
);
CREATE TABLE ThreadNames (
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the thread name.
    priority                    INTEGER,                               -- Priority of the thread.
    globalTid                   INTEGER                                -- Serialized GlobalId.
);
CREATE TABLE ProcessStreams (
    globalPid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    filenameId                  INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the file name.
    contentId                   INTEGER   NOT NULL   REFERENCES StringIds(id) -- StringId of the stream content.
);
CREATE TABLE TARGET_INFO_SYSTEM_ENV (
    globalVid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    devStateName                TEXT      NOT NULL,                    -- Device state name.
    name                        TEXT      NOT NULL,                    -- Property name.
    nameEnum                    INTEGER   NOT NULL,                    -- Property enum value.
    value                       TEXT      NOT NULL                     -- Property value.
);
CREATE TABLE TARGET_INFO_SESSION_START_TIME (
    utcEpochNs                  INTEGER,                               -- UTC Epoch timestamp at start of the capture (ns).
    utcTime                     TEXT,                                  -- Start of the capture in UTC.
    localTime                   TEXT                                   -- Start of the capture in local time of target.
);
CREATE TABLE ANALYSIS_DETAILS (
    -- Details about the analysis session.

    globalVid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    duration                    INTEGER   NOT NULL,                    -- The total time span of the entire trace (ns).
    startTime                   INTEGER   NOT NULL,                    -- Trace start timestamp in nanoseconds.
    stopTime                    INTEGER   NOT NULL                     -- Trace stop timestamp in nanoseconds.
);
CREATE TABLE TARGET_INFO_GPU (
    vmId                        INTEGER   NOT NULL,                    -- Serialized GlobalId.
    id                          INTEGER   NOT NULL,                    -- Device ID.
    name                        TEXT,                                  -- Device name.
    busLocation                 TEXT,                                  -- PCI bus location.
    isDiscrete                  INTEGER,                               -- True if discrete, false if integrated.
    l2CacheSize                 INTEGER,                               -- Size of L2 cache (B).
    totalMemory                 INTEGER,                               -- Total amount of memory on the device (B).
    memoryBandwidth             INTEGER,                               -- Amount of memory transferred (B).
    clockRate                   INTEGER,                               -- Clock frequency (Hz).
    smCount                     INTEGER,                               -- Number of multiprocessors on the device.
    pwGpuId                     INTEGER,                               -- PerfWorks GPU ID.
    uuid                        TEXT,                                  -- Device UUID.
    luid                        INTEGER,                               -- Device LUID.
    chipName                    TEXT,                                  -- Chip name.
    cuDevice                    INTEGER,                               -- CUDA device ID.
    ctxswDevPath                TEXT,                                  -- GPU context switch device node path.
    ctrlDevPath                 TEXT,                                  -- GPU control device node path.
    revision                    INTEGER,                               -- Revision number.
    nodeMask                    INTEGER,                               -- Device node mask.
    constantMemory              INTEGER,                               -- Memory available on device for __constant__ variables (B).
    maxIPC                      INTEGER,                               -- Maximum instructions per count.
    maxRegistersPerBlock        INTEGER,                               -- Maximum number of 32-bit registers available per block.
    maxShmemPerBlock            INTEGER,                               -- Maximum optin shared memory per block.
    maxShmemPerBlockOptin       INTEGER,                               -- Maximum optin shared memory per block.
    maxShmemPerSm               INTEGER,                               -- Maximum shared memory available per multiprocessor (B).
    maxRegistersPerSm           INTEGER,                               -- Maximum number of 32-bit registers available per multiprocessor.
    threadsPerWarp              INTEGER,                               -- Warp size in threads.
    asyncEngines                INTEGER,                               -- Number of asynchronous engines.
    maxWarpsPerSm               INTEGER,                               -- Maximum number of warps per multiprocessor.
    maxBlocksPerSm              INTEGER,                               -- Maximum number of blocks per multiprocessor.
    maxThreadsPerBlock          INTEGER,                               -- Maximum number of threads per block.
    maxBlockDimX                INTEGER,                               -- Maximum X-dimension of a block.
    maxBlockDimY                INTEGER,                               -- Maximum Y-dimension of a block.
    maxBlockDimZ                INTEGER,                               -- Maximum Z-dimension of a block.
    maxGridDimX                 INTEGER,                               -- Maximum X-dimension of a grid.
    maxGridDimY                 INTEGER,                               -- Maximum Y-dimension of a grid.
    maxGridDimZ                 INTEGER,                               -- Maximum Z-dimension of a grid.
    computeMajor                INTEGER,                               -- Major compute capability version number.
    computeMinor                INTEGER,                               -- Minor compute capability version number.
    smMajor                     INTEGER,                               -- Major multiprocessor version number.
    smMinor                     INTEGER                                -- Minor multiprocessor version number.
);
CREATE TABLE TARGET_INFO_XMC_SPEC (
    vmId                        INTEGER   NOT NULL,                    -- Serialized GlobalId.
    clientId                    INTEGER   NOT NULL,                    -- Client ID.
    type                        TEXT      NOT NULL,                    -- Client type.
    name                        TEXT      NOT NULL,                    -- Client name.
    groupId                     TEXT      NOT NULL                     -- Client group ID.
);
CREATE TABLE TARGET_INFO_PROCESS (
    processId                   INTEGER   NOT NULL,                    -- Process ID.
    openGlVersion               TEXT      NOT NULL,                    -- OpenGL version.
    correlationId               INTEGER   NOT NULL,                    -- Correlation ID of the kernel.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id) -- StringId of the function name.
);
CREATE TABLE TARGET_INFO_NVTX_CUDA_DEVICE (
    name                        TEXT      NOT NULL,                    -- CUDA device name assigned using NVTX.
    hwId                        INTEGER   NOT NULL,                    -- Hardware ID.
    vmId                        INTEGER   NOT NULL,                    -- VM ID.
    deviceId                    INTEGER   NOT NULL                     -- Device ID.
);
CREATE TABLE TARGET_INFO_NVTX_CUDA_CONTEXT (
    name                        TEXT      NOT NULL,                    -- CUDA context name assigned using NVTX.
    hwId                        INTEGER   NOT NULL,                    -- Hardware ID.
    vmId                        INTEGER   NOT NULL,                    -- VM ID.
    processId                   INTEGER   NOT NULL,                    -- Process ID.
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL                     -- Context ID.
);
CREATE TABLE TARGET_INFO_NVTX_CUDA_STREAM (
    name                        TEXT      NOT NULL,                    -- CUDA stream name assigned using NVTX.
    hwId                        INTEGER   NOT NULL,                    -- Hardware ID.
    vmId                        INTEGER   NOT NULL,                    -- VM ID.
    processId                   INTEGER   NOT NULL,                    -- Process ID.
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    streamId                    INTEGER   NOT NULL                     -- Stream ID.
);
CREATE TABLE TARGET_INFO_CUDA_NULL_STREAM (
    streamId                    INTEGER   NOT NULL,                    -- Stream ID.
    hwId                        INTEGER   NOT NULL,                    -- Hardware ID.
    vmId                        INTEGER   NOT NULL,                    -- VM ID.
    processId                   INTEGER   NOT NULL,                    -- Process ID.
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL                     -- Context ID.
);
CREATE TABLE TARGET_INFO_CUDA_STREAM (
    streamId                    INTEGER   NOT NULL,                    -- Stream ID.
    hwId                        INTEGER   NOT NULL,                    -- Hardware ID.
    vmId                        INTEGER   NOT NULL,                    -- VM ID.
    processId                   INTEGER   NOT NULL,                    -- Process ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    priority                    INTEGER   NOT NULL,                    -- Priority of the stream.
    flag                        INTEGER   NOT NULL                     -- Flags for stream creation.
);
CREATE TABLE TARGET_INFO_WDDM_CONTEXTS (
    context                     INTEGER   NOT NULL,
    engineType                  INTEGER   NOT NULL,
    nodeOrdinal                 INTEGER   NOT NULL,
    friendlyName                TEXT      NOT NULL
);
CREATE TABLE EXPORT_META_DATA (
    -- information about nsys export process

    name                        TEXT      NOT NULL,                    -- Name of export meta-data record
    value                       TEXT                                   -- Value of export meta-data record
);
CREATE TABLE CUPTI_ACTIVITY_KIND_MEMCPY (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    streamId                    INTEGER   NOT NULL,                    -- Stream ID.
    correlationId               INTEGER   REFERENCES CUPTI_ACTIVITY_KIND_RUNTIME(correlationId), -- ID used to identify the API call that launched the event.
    globalPid                   INTEGER,                               -- Serialized GlobalId.
    bytes                       INTEGER   NOT NULL,                    -- Number of bytes transferred (B).
    copyKind                    INTEGER   NOT NULL,                    -- Type of the memory copy. See enum CUDA_MEMCPY_KIND.
    deprecatedSrcId             INTEGER,                               -- Deprecated, use srcDeviceId instead.
    srcKind                     INTEGER,                               -- Source memory kind See enum CUDA_MEMOPR_MEMORY_KIND.
    dstKind                     INTEGER,                               -- Destination memory kind. See enum CUDA_MEMOPR_MEMORY_KIND.
    srcDeviceId                 INTEGER,                               -- Source device ID.
    srcContextId                INTEGER,                               -- Source context ID.
    dstDeviceId                 INTEGER,                               -- Destination device ID.
    dstContextId                INTEGER,                               -- Destination context ID.
    migrationCause              INTEGER,                               -- UVM migration cause ID. See CUpti_ActivityUnifiedMemoryMigrationCause.
    graphNodeId                 INTEGER   REFERENCES CUDA_GRAPH_EVENTS(graphNodeId), -- CUDA graph node creation ID.
    virtualAddress              INTEGER                                -- Virtual base address of the page/s being transferred.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_MEMSET (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    streamId                    INTEGER   NOT NULL,                    -- Stream ID.
    correlationId               INTEGER   REFERENCES CUPTI_ACTIVITY_KIND_RUNTIME(correlationId), -- ID used to identify the API call that launched the event.
    globalPid                   INTEGER,                               -- Serialized GlobalId.
    value                       INTEGER   NOT NULL,                    -- Value assigned to memory.
    bytes                       INTEGER   NOT NULL,                    -- Number of bytes set (B).
    graphNodeId                 INTEGER   REFERENCES CUDA_GRAPH_EVENTS(graphNodeId), -- CUDA graph node creation ID.
    memKind                     INTEGER                                -- Type of memory being set. See enum CUDA_MEMOPR_MEMORY_KIND.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_KERNEL (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    streamId                    INTEGER   NOT NULL,                    -- Stream ID.
    correlationId               INTEGER   REFERENCES CUPTI_ACTIVITY_KIND_RUNTIME(correlationId), -- ID used to identify the API call that launched the event.
    globalPid                   INTEGER,                               -- Serialized GlobalId.
    demangledName               INTEGER   NOT NULL,                    -- StringId of the kernel name
    shortName                   INTEGER   NOT NULL,                    -- StringId of the short function name without arguments.
    mangledName                 INTEGER,                               -- StringId of the mangled function name.
    launchType                  INTEGER,                               -- Launch type of the kernel. See docs for specifics.
    cacheConfig                 INTEGER,                               -- Cache configuration used for the kernel. See enum CUfunc_cache.
    registersPerThread          INTEGER   NOT NULL,                    -- Number of registers required for each thread executing the kernel.
    gridX                       INTEGER   NOT NULL,                    -- X-dimension grid size.
    gridY                       INTEGER   NOT NULL,                    -- Y-dimension grid size.
    gridZ                       INTEGER   NOT NULL,                    -- Z-dimension grid size.
    blockX                      INTEGER   NOT NULL,                    -- X-dimension block size.
    blockY                      INTEGER   NOT NULL,                    -- Y-dimension block size.
    blockZ                      INTEGER   NOT NULL,                    -- Z-dimension block size.
    staticSharedMemory          INTEGER   NOT NULL,                    -- Static shared memory allocated for the kernel (B).
    dynamicSharedMemory         INTEGER   NOT NULL,                    -- Dynamic shared memory reserved for the kernel (B).
    localMemoryPerThread        INTEGER   NOT NULL,                    -- Amount of local memory reserved for each thread (B).
    localMemoryTotal            INTEGER   NOT NULL,                    -- Total amount of local memory reserved for the kernel (B).
    gridId                      INTEGER   NOT NULL,                    -- Unique grid ID of the kernel assigned at runtime.
    sharedMemoryExecuted        INTEGER,                               -- Shared memory size set by the driver.
    graphNodeId                 INTEGER   REFERENCES CUDA_GRAPH_EVENTS(graphNodeId), -- CUDA graph node creation ID.
    sharedMemoryLimitConfig     INTEGER                                -- Shared memory limit per block config. See enum CUpti_FuncShmemLimitConfig.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_CUDA_EVENT (
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    streamId                    INTEGER   NOT NULL,                    -- Stream ID.
    correlationId               INTEGER,                               -- Correlation ID of the event record API to which this result is associated.
    globalPid                   INTEGER,                               -- Serialized GlobalId.
    eventId                     INTEGER   NOT NULL                     -- Event ID for which the event record API is called.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_SYNCHRONIZATION (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    streamId                    INTEGER   NOT NULL,                    -- Stream ID.
    correlationId               INTEGER,                               -- Correlation ID of the API to which this result is associated.
    globalPid                   INTEGER,                               -- Serialized GlobalId.
    syncType                    INTEGER   NOT NULL,                    -- Type of synchronization. See enum CUpti_ActivitySynchronizationType.
    eventId                     INTEGER   NOT NULL                     -- Event ID for which the synchronization API is called.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_RUNTIME (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- CUDA event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- ID used to identify events that this function call has triggered.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    returnValue                 INTEGER   NOT NULL,                    -- Return value of the function call.
    callchainId                 INTEGER   REFERENCES CUDA_CALLCHAINS(id) -- ID of the attached callchain.
);
CREATE TABLE CUDNN_EVENTS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- CUDA event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id) -- StringId of the function name.
);
CREATE TABLE CUBLAS_EVENTS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- CUDA event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id) -- StringId of the function name.
);
CREATE TABLE CUDA_GRAPH_EVENTS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- CUDA event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    graphNodeId                 INTEGER   NOT NULL   REFERENCES CUDA_GRAPH_EVENTS(graphNodeId), -- CUDA graph node creation ID.
    originalGraphNodeId         INTEGER                                -- Reference to the original graph node ID, if cloned node.
);
CREATE TABLE CUDA_UM_CPU_PAGE_FAULT_EVENTS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    globalPid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    address                     INTEGER   NOT NULL,                    -- Virtual address of the page that faulted.
    originalFaultPc             INTEGER,                               -- Program counter of the CPU instruction that caused the page fault.
    CpuInstruction              INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    module                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the module name.
    unresolvedFaultPc           INTEGER                                -- True if the program counter was not resolved.
);
CREATE TABLE CUDA_UM_GPU_PAGE_FAULT_EVENTS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalPid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    address                     INTEGER   NOT NULL,                    -- Virtual address of the page that faulted.
    numberOfPageFaults          INTEGER   NOT NULL,                    -- Number of page faults for the same page.
    faultAccessType             INTEGER   NOT NULL                     -- Type of Unified memory accessed. See enum CUpti_ActivityUnifiedMemoryAccessType.
);
CREATE TABLE CUDA_GPU_MEMORY_USAGE_EVENTS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    globalPid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    deviceId                    INTEGER   NOT NULL,                    -- Device ID.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    address                     INTEGER   NOT NULL,                    -- Virtual address of the allocation/deallocation.
    pc                          INTEGER   NOT NULL,                    -- Program counter of the allocation/deallocation.
    bytes                       INTEGER   NOT NULL,                    -- Number of bytes allocated/deallocated (B).
    memKind                     INTEGER   NOT NULL,                    -- Type of memory being allocated/deallocated. See enum CUDA_MEMOPR_MEMORY_KIND.
    memoryOperationType         INTEGER   NOT NULL,                    -- 0 if memory was allocated and 1 if memory was deallocated.
    name                        TEXT,                                  -- Variable name, if available.
    correlationId               INTEGER                                -- ID used to identify the API call that launched the event.
);
CREATE TABLE CUDA_CALLCHAINS (
    id                          INTEGER   NOT NULL,                    -- Part of PRIMARY KEY (id, stackDepth).
    symbol                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    module                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the module name.
    unresolved                  INTEGER,                               -- True if the symbol was not resolved.
    originalIP                  INTEGER,                               -- Instruction pointer value.
    stackDepth                  INTEGER   NOT NULL,                    -- Zero-base index of the given function in call stack.

    PRIMARY KEY (id, stackDepth)
);
CREATE TABLE NVTX_EVENTS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER,                               -- Event end timestamp (ns).
    eventType                   INTEGER   NOT NULL,                    -- NVTX event type enum value. See docs for specifics.
    rangeId                     INTEGER,                               -- Correlation ID returned from a nvtxRangeStart call.
    category                    INTEGER,                               -- User-controlled ID that can be used to group events.
    color                       INTEGER,                               -- Encoded ARGB color value.
    text                        TEXT,                                  -- Optional text message for non registered strings.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    endGlobalTid                INTEGER,                               -- Serialized GlobalId. See docs for specifics.
    textId                      INTEGER   REFERENCES StringIds(id),    -- StringId of the NVTX domain registered string.
    domainId                    INTEGER,                               -- User-controlled ID that can be used to group events.
    uint64Value                 INTEGER,                               -- One of possible payload value union members.
    int64Value                  INTEGER,                               -- One of possible payload value union members.
    doubleValue                 REAL,                                  -- One of possible payload value union members.
    uint32Value                 INTEGER,                               -- One of possible payload value union members.
    int32Value                  INTEGER,                               -- One of possible payload value union members.
    floatValue                  REAL,                                  -- One of possible payload value union members.
    jsonTextId                  INTEGER,                               -- One of possible payload value union members.
    jsonText                    TEXT                                   -- One of possible payload value union members.
);
CREATE TABLE OPENGL_API (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenGL event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    endGlobalTid                INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the first function name.
    endNameId                   INTEGER   REFERENCES StringIds(id),    -- StringId of the last function name.
    returnValue                 INTEGER   NOT NULL,                    -- Return value of the function call.
    frameId                     INTEGER,                               -- Index of the graphics frame starting from 1.
    contextId                   INTEGER,                               -- Context ID.
    gpu                         INTEGER,                               -- GPU index.
    display                     INTEGER                                -- Display ID.
);
CREATE TABLE OPENGL_WORKLOAD (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenGL event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    endGlobalTid                INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the first function name.
    endNameId                   INTEGER   REFERENCES StringIds(id),    -- StringId of the last function name.
    returnValue                 INTEGER   NOT NULL,                    -- Return value of the function call.
    frameId                     INTEGER,                               -- Index of the graphics frame starting from 1.
    contextId                   INTEGER,                               -- Context ID.
    gpu                         INTEGER,                               -- GPU index.
    display                     INTEGER                                -- Display ID.
);
CREATE TABLE KHR_DEBUG_EVENTS (
    eventClass                  INTEGER   NOT NULL,                    -- KHR event class enum value. See docs for specifics.
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER,                               -- Event end timestamp (ns).
    textId                      INTEGER   REFERENCES StringIds(id),    -- StringId of the debug marker/group text string.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    source                      INTEGER,                               -- KHR source enum value. See docs for specifics.
    khrdType                    INTEGER,                               -- KHR type enum value. See docs for specifics.
    id                          INTEGER,                               -- KHR event ID.
    severity                    INTEGER,                               -- KHR severity enum value. See docs for specifics.
    correlationId               INTEGER,                               -- ID used to correlate KHR CPU trace to GPU trace.
    context                     INTEGER                                -- Context ID.
);
CREATE TABLE OSRT_API (
    -- OS runtime libraries traced to gather information about low-level userspace APIs.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OSRT event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    returnValue                 INTEGER   NOT NULL,                    -- Return value of the function call.
    nestingLevel                INTEGER,                               -- Zero-base index of the nesting level.
    callchainId                 INTEGER   NOT NULL   REFERENCES OSRT_CALLCHAINS(id) -- ID of the attached callchain.
);
CREATE TABLE OSRT_CALLCHAINS (
    -- Callchains attached to OSRT events, depending on selected profiling settings.

    id                          INTEGER   NOT NULL,                    -- Part of PRIMARY KEY (id, stackDepth).
    symbol                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    module                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the module name.
    kernelMode                  INTEGER,                               -- True if kernel mode.
    thumbCode                   INTEGER,                               -- True if thumb code.
    unresolved                  INTEGER,                               -- True if the symbol was not resolved.
    specialEntry                INTEGER,                               -- True if artifical entry added during processing callchain.
    originalIP                  INTEGER,                               -- Instruction pointer value.
    unwindMethod                INTEGER   REFERENCES UnwindMethodType(number), -- Instruction pointer value.
    stackDepth                  INTEGER   NOT NULL,                    -- Zero-base index of the given function in call stack.

    PRIMARY KEY (id, stackDepth)
);
CREATE TABLE UnwindMethodType (
    number                      INTEGER   PRIMARY KEY,
    name                        TEXT      NOT NULL
);
CREATE TABLE PROFILER_OVERHEAD (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    returnValue                 INTEGER   NOT NULL                     -- Return value of the function call.
);
CREATE TABLE SCHED_EVENTS (
    -- Thread scheduling events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    cpu                         INTEGER   NOT NULL,                    -- ID of CPU this thread was scheduled in or out.
    isSchedIn                   INTEGER   NOT NULL,                    -- 0 if thread was scheduled out, non-zero otherwise.
    globalTid                   INTEGER                                -- Serialized GlobalId.
);
CREATE TABLE COMPOSITE_EVENTS (
    -- Thread sampling events.

    id                          INTEGER   NOT NULL   PRIMARY KEY,      -- ID of the composite event.
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    cpu                         INTEGER,                               -- ID of CPU this thread was running on.
    threadState                 INTEGER,                               -- Thread state at the moment of sampling. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    cpuCycles                   INTEGER   NOT NULL                     -- Value of Performance Monitoring Unit (PMU) counter.
);
CREATE TABLE SAMPLING_CALLCHAINS (
    -- Callchain entries obtained from composite events, used to construct function table views.

    id                          INTEGER   NOT NULL   REFERENCES COMPOSITE_EVENTS(id), -- ID of the attached composite event.
    symbol                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    module                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the module name.
    kernelMode                  INTEGER,                               -- True if kernel mode.
    thumbCode                   INTEGER,                               -- True if thumb code.
    unresolved                  INTEGER,                               -- True if the symbol was not resolved.
    specialEntry                INTEGER,                               -- True if artifical entry added during processing callchain.
    originalIP                  INTEGER,                               -- Instruction pointer value.
    unwindMethod                INTEGER   REFERENCES UnwindMethodType(number), -- Instruction pointer value.
    stackDepth                  INTEGER   NOT NULL,                    -- Zero-base index of the given function in call stack.

    PRIMARY KEY (id, stackDepth)
);
CREATE TABLE SLI_QUERIES (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- SLI event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    frameId                     INTEGER   NOT NULL,                    -- Index of the graphics frame starting from 1.
    occQueryIssued              INTEGER   NOT NULL,                    -- Occlusion query issued.
    occQueryAsked               INTEGER   NOT NULL,                    -- Occlusion query asked.
    eventQueryIssued            INTEGER   NOT NULL,                    -- Event query issued.
    eventQueryAsked             INTEGER   NOT NULL,                    -- Event query asked.
    numberOfTransferEvents      INTEGER   NOT NULL,                    -- Number of transfer events.
    amountOfTransferredData     INTEGER   NOT NULL                     -- Cumulative size of resource data that was transferred.
);
CREATE TABLE SLI_P2P (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- SLI event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    frameId                     INTEGER   NOT NULL,                    -- Index of the graphics frame starting from 1.
    transferSkipped             INTEGER   NOT NULL,                    -- Number of transfers that were skipped.
    srcGpu                      INTEGER   NOT NULL,                    -- Source GPU ID.
    dstGpu                      INTEGER   NOT NULL,                    -- Destination GPU ID.
    numSubResources             INTEGER   NOT NULL,                    -- Number of sub-resources to transfer.
    resourceSize                INTEGER   NOT NULL,                    -- Size of resource.
    subResourceIdx              INTEGER   NOT NULL,                    -- Sub-resource index.
    smplWidth                   INTEGER,                               -- Sub-resource surface width in samples.
    smplHeight                  INTEGER,                               -- Sub-resource surface height in samples.
    smplDepth                   INTEGER,                               -- Sub-resource surface depth in samples.
    bytesPerElement             INTEGER,                               -- Number of bytes per element.
    dxgiFormat                  INTEGER,                               -- Resource data format. See DXGI_FORMAT enum.
    logSurfaceNames             TEXT,                                  -- Surface name.
    transferInfo                INTEGER,                               -- SLI transfer info enum value.  See docs for specifics.
    isEarlyPushManagedByNvApi   INTEGER,                               -- True if early push managed by NVAPI. False otherwise.
    useAsyncP2pForResolve       INTEGER,                               -- True if async Peer-to-Peer used for resolve. False otherwise.
    transferFuncName            TEXT,                                  -- "A - BE" for asynchronous transfer, "S - BE" for synchronous transfer.
    regimeName                  TEXT,                                  -- Name of the regime scope that includes the resource.
    debugName                   TEXT,                                  -- Debug name assigned to the resource by the application code.
    bindType                    TEXT                                   -- Bind type.
);
CREATE TABLE SLI_STATS (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- SLI event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    countComplexFrames          INTEGER   NOT NULL,                    -- Complex frames count.
    countStats                  INTEGER   NOT NULL,                    -- Number of frame statistics collected for the inactive-time histogram.
    totalInactiveTime           INTEGER   NOT NULL,                    -- Total inactive time (μs).
    minPbSize                   INTEGER   NOT NULL,                    -- Min push buffer size.
    maxPbSize                   INTEGER   NOT NULL,                    -- Max push buffer size.
    totalPbSize                 INTEGER   NOT NULL                     -- Total push buffer size.
);
CREATE TABLE DX12_API (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- DX12 event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    shortContextId              INTEGER,                               -- Short form of the COM interface object address.
    frameId                     INTEGER,                               -- Index of the graphics frame starting from 1.
    color                       INTEGER,                               -- Encoded ARGB color value.
    textId                      INTEGER,                               -- StringId of the PIX marker text string.
    commandListType             INTEGER,                               -- Type of command list. See enum D3D12_COMMAND_LIST_TYPE.
    objectName                  TEXT,                                  -- StringId of the D3D12 object name.
    longContextId               INTEGER                                -- Long form of the COM interface object address.
);
CREATE TABLE DX12_WORKLOAD (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- DX12 event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    shortContextId              INTEGER,                               -- Short form of the COM interface object address.
    frameId                     INTEGER,                               -- Index of the graphics frame starting from 1.
    gpu                         INTEGER,                               -- GPU index.
    color                       INTEGER,                               -- Encoded ARGB color value.
    textId                      INTEGER,                               -- StringId of the PIX marker text string.
    commandListType             INTEGER,                               -- Type of command list. See enum D3D12_COMMAND_LIST_TYPE.
    objectName                  TEXT,                                  -- StringId of the D3D12 object name.
    longContextId               INTEGER                                -- Long form of the COM interface object address.
);
CREATE TABLE VULKAN_API (
    id                          INTEGER   NOT NULL   PRIMARY KEY,
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- Vulkan event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    contextId                   INTEGER                                -- Short form of the interface object address.
);
CREATE TABLE VULKAN_WORKLOAD (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- Vulkan event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    gpu                         INTEGER,                               -- GPU index.
    contextId                   INTEGER,                               -- Short form of the interface object address.
    color                       INTEGER,                               -- Encoded ARGB color value.
    textId                      INTEGER                                -- StringId of the VULKAN CPU debug marker text string.
);
CREATE TABLE VULKAN_DEBUG_API (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- Vulkan event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    contextId                   INTEGER,                               -- Short form of the interface object address.
    color                       INTEGER,                               -- Encoded ARGB color value.
    textId                      INTEGER                                -- StringId of the VULKAN CPU debug marker text string.
);
CREATE TABLE VULKAN_PIPELINE_CREATION_EVENTS (
    id                          INTEGER   NOT NULL   PRIMARY KEY,      -- ID of the pipeline creation event.
    duration                    INTEGER,                               -- Event duration (ns).
    flags                       INTEGER,
    traceEventId                INTEGER   NOT NULL   REFERENCES VULKAN_API(id) -- ID of the attached vulkan API.
);
CREATE TABLE VULKAN_PIPELINE_STAGE_EVENTS (
    id                          INTEGER   NOT NULL   PRIMARY KEY,      -- ID of the pipeline stage event.
    duration                    INTEGER,                               -- Event duration (ns).
    flags                       INTEGER,                               -- Vulkan flag enum value. See docs for specifics.
    creationEventId             INTEGER   NOT NULL   REFERENCES VULKAN_PIPELINE_CREATION_EVENTS(id) -- ID of the attached pipeline creation event.
);
CREATE TABLE FECS_EVENTS (
    tag                         INTEGER   NOT NULL,                    -- Context switch event type. See docs for specifics.
    vmId                        INTEGER   NOT NULL,                    -- VM ID.
    seqNo                       INTEGER   NOT NULL,                    -- Sequential event number.
    contextId                   INTEGER   NOT NULL,                    -- Context ID.
    timestamp                   INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    globalPid                   INTEGER,                               -- Serialized GlobalId.
    gpuId                       INTEGER                                -- GPU index.
);
CREATE TABLE ETW_EVENTS_DEPRECATED_TABLE (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    providerName                INTEGER   REFERENCES StringIds(id),    -- StringId of tracing events provider name.
    taskName                    INTEGER   REFERENCES StringIds(id),    -- StringId of the event task name.
    description                 INTEGER   REFERENCES StringIds(id)     -- StringId of the decoded event value.
);
CREATE TABLE ETW_PROVIDERS (
    -- Names and identifiers of ETW providers captured in the report.

    providerId                  INTEGER   NOT NULL   PRIMARY KEY,      -- Provider ID.
    providerNameId              INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of provider name.
    guid                        TEXT      NOT NULL                     -- ETW Provider GUID.
);
CREATE TABLE ETW_TASKS (
    -- Names and identifiers of ETW tasks captured in the report.

    taskNameId                  INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of task name.
    taskId                      INTEGER   NOT NULL,                    -- The event task ID.
    providerId                  INTEGER   NOT NULL                     -- Provider ID.
);
CREATE TABLE ETW_EVENTS (
    -- Raw ETW events captured in the report.

    timestamp                   INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    processId                   INTEGER,                               -- Process ID.
    threadId                    INTEGER,                               -- Thread ID.
    providerId                  INTEGER   NOT NULL,                    -- Provider ID.
    taskId                      INTEGER   NOT NULL,                    -- The event task ID.
    eventId                     INTEGER   NOT NULL,                    -- Event ID.
    version                     INTEGER   NOT NULL,                    -- The event version.
    opcode                      INTEGER,                               -- The event opcode.
    data                        TEXT      NOT NULL                     -- JSON encoded event data.
);
CREATE TABLE OPENMP_EVENT_KIND_THREAD (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    threadId                    INTEGER,                               -- Internal thread sequence starting from 1.
    threadType                  INTEGER                                -- See enum ompt_thread_t.
);
CREATE TABLE OPENMP_EVENT_KIND_PARALLEL (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parallelId                  INTEGER,                               -- Internal parallel region sequence starting from 1.
    parentTaskId                INTEGER                                -- ID for task that creates this parallel region.
);
CREATE TABLE OPENMP_EVENT_KIND_SYNC_REGION_WAIT (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    taskId                      INTEGER,                               -- ID of the task that this event belongs to.
    kind                        INTEGER                                -- See enum ompt_sync_region_t.
);
CREATE TABLE OPENMP_EVENT_KIND_SYNC_REGION (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    taskId                      INTEGER,                               -- ID of the task that this event belongs to.
    kind                        INTEGER                                -- See enum ompt_sync_region_t.
);
CREATE TABLE OPENMP_EVENT_KIND_TASK (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    taskId                      INTEGER,                               -- ID of the task that this event belongs to.
    kind                        INTEGER                                -- See enum ompt_task_flag_t.
);
CREATE TABLE OPENMP_EVENT_KIND_MASTER (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    taskId                      INTEGER                                -- ID of the task that this event belongs to.
);
CREATE TABLE OPENMP_EVENT_KIND_REDUCTION (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    taskId                      INTEGER                                -- ID of the task that this event belongs to.
);
CREATE TABLE OPENMP_EVENT_KIND_TASK_CREATE (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parentTaskId                INTEGER,                               -- ID of the parent task that is creating a new task.
    newTaskId                   INTEGER                                -- ID of the new task that is being created.
);
CREATE TABLE OPENMP_EVENT_KIND_TASK_SCHEDULE (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    priorTaskId                 INTEGER,                               -- ID of the task that is being switched out.
    priorTaskStatus             INTEGER,                               -- See enum ompt_task_status_t.
    nextTaskId                  INTEGER                                -- ID of the task that is being switched in.
);
CREATE TABLE OPENMP_EVENT_KIND_CANCEL (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    taskId                      INTEGER                                -- ID of the task that is being cancelled.
);
CREATE TABLE OPENMP_EVENT_KIND_MUTEX_WAIT (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    kind                        INTEGER,                               -- See enum ompt_mutex_t.
    waitId                      INTEGER,                               -- ID indicating the object being waited.
    taskId                      INTEGER                                -- ID of the task that this event belongs to.
);
CREATE TABLE OPENMP_EVENT_KIND_CRITICAL_SECTION (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    kind                        INTEGER,                               -- See enum ompt_mutex_t.
    waitId                      INTEGER                                -- ID indicating the object being held.
);
CREATE TABLE OPENMP_EVENT_KIND_MUTEX_RELEASED (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    kind                        INTEGER,                               -- See enum ompt_mutex_t.
    waitId                      INTEGER,                               -- ID indicating the object being released.
    taskId                      INTEGER                                -- ID of the task that this event belongs to.
);
CREATE TABLE OPENMP_EVENT_KIND_LOCK_INIT (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    kind                        INTEGER,                               -- See enum ompt_mutex_t.
    waitId                      INTEGER                                -- ID indicating object being created/destroyed.
);
CREATE TABLE OPENMP_EVENT_KIND_LOCK_DESTROY (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    kind                        INTEGER,                               -- See enum ompt_mutex_t.
    waitId                      INTEGER                                -- ID indicating object being created/destroyed.
);
CREATE TABLE OPENMP_EVENT_KIND_WORKSHARE (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    kind                        INTEGER,                               -- See enum ompt_work_t.
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    taskId                      INTEGER,                               -- ID of the task that this event belongs to.
    count                       INTEGER                                -- Measure of the quantity of work involved in the region.
);
CREATE TABLE OPENMP_EVENT_KIND_DISPATCH (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    kind                        INTEGER,                               -- See enum ompt_dispatch_t
    parallelId                  INTEGER,                               -- ID of the parallel region that this event belongs to.
    taskId                      INTEGER                                -- ID of the task that this event belongs to.
);
CREATE TABLE OPENMP_EVENT_KIND_FLUSH (
    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- OpenMP event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- Currently unused.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    eventKind                   INTEGER,                               -- OpenMP event kind enum value. See docs for specifics.
    threadId                    INTEGER                                -- ID of the thread that this event belongs to.
);
CREATE TABLE D3D11_PIX_DEBUG_API (
    -- D3D11 debug marker events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    shortContextId              INTEGER,                               -- Short form of the COM interface object address.
    frameId                     INTEGER,                               -- Index of the graphics frame starting from 1.
    color                       INTEGER,                               -- Encoded ARGB color value.
    textId                      INTEGER                                -- StringId of the PIX marker text string.
);
CREATE TABLE D3D12_PIX_DEBUG_API (
    -- D3D12 debug marker events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    correlationId               INTEGER,                               -- First ID matching an API call to GPU workloads.
    endCorrelationId            INTEGER,                               -- Last ID matching an API call to GPU workloads.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the function name.
    shortContextId              INTEGER,                               -- Short form of the COM interface object address.
    frameId                     INTEGER,                               -- Index of the graphics frame starting from 1.
    color                       INTEGER,                               -- Encoded ARGB color value.
    textId                      INTEGER,                               -- StringId of the PIX marker text string.
    commandListType             INTEGER,                               -- Type of command list. See enum D3D12_COMMAND_LIST_TYPE.
    objectName                  TEXT,                                  -- StringId of the D3D12 object name.
    longContextId               INTEGER                                -- Long form of the COM interface object address.
);
CREATE TABLE WDDM_EVICT_ALLOCATION_EVENTS (
    -- Raw ETW EvictAllocation events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    allocationHandle            INTEGER   NOT NULL                     -- Global allocation handle.
);
CREATE TABLE WDDM_PAGING_QUEUE_PACKET_START_EVENTS (
    -- Raw ETW PagingQueuePacketStart events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    dxgDevice                   INTEGER,                               -- Address of an IDXGIDevice.
    dxgAdapter                  INTEGER,                               -- Address of an IDXGIAdapter.
    pagingQueue                 INTEGER   NOT NULL,                    -- Address of the paging queue.
    pagingQueuePacket           INTEGER   NOT NULL,                    -- Address of the paging queue packet.
    sequenceId                  INTEGER   NOT NULL,                    -- Internal sequence starting from 0.
    alloc                       INTEGER,                               -- Allocation handle.
    vidMmOpType                 INTEGER   NOT NULL,                    -- VIDMM operation type enum value. See docs for specifics.
    pagingQueueType             INTEGER   NOT NULL                     -- Paging queue type enum value. See docs for specifics.
);
CREATE TABLE WDDM_PAGING_QUEUE_PACKET_STOP_EVENTS (
    -- Raw ETW PagingQueuePacketStop events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    pagingQueue                 INTEGER   NOT NULL,                    -- Address of the paging queue.
    pagingQueuePacket           INTEGER   NOT NULL,                    -- Address of the paging queue packet.
    sequenceId                  INTEGER   NOT NULL                     -- Internal sequence starting from 0.
);
CREATE TABLE WDDM_PAGING_QUEUE_PACKET_INFO_EVENTS (
    -- Raw ETW PagingQueuePacketInfo events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    pagingQueue                 INTEGER   NOT NULL,                    -- Address of the paging queue.
    pagingQueuePacket           INTEGER   NOT NULL,                    -- Address of the paging queue packet.
    sequenceId                  INTEGER   NOT NULL                     -- Internal sequence starting from 0.
);
CREATE TABLE WDDM_QUEUE_PACKET_START_EVENTS (
    -- Raw ETW QueuePacketStart events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    context                     INTEGER   NOT NULL,                    -- The context ID of WDDM queue.
    dmaBufferSize               INTEGER   NOT NULL,                    -- The dma buffer size.
    dmaBuffer                   INTEGER   NOT NULL,                    -- The reported address of dma buffer.
    queuePacket                 INTEGER   NOT NULL,                    -- The address of queue packet.
    progressFenceValue          INTEGER   NOT NULL,                    -- The fence value.
    packetType                  INTEGER   NOT NULL,                    -- The type of the packet.
    submitSequence              INTEGER   NOT NULL,                    -- Internal sequence starting from 1.
    allocationListSize          INTEGER   NOT NULL,                    -- The number of allocations referenced.
    patchLocationListSize       INTEGER   NOT NULL,                    -- The number of patch locations.
    present                     INTEGER   NOT NULL,                    -- True or False if the packet is a present packet.
    engineType                  INTEGER   NOT NULL,                    -- The engine type.
    syncObject                  INTEGER                                -- The address of fence object.
);
CREATE TABLE WDDM_QUEUE_PACKET_STOP_EVENTS (
    -- Raw ETW QueuePacketStop events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    context                     INTEGER   NOT NULL,                    -- The context ID of WDDM queue.
    queuePacket                 INTEGER   NOT NULL,                    -- The address of queue packet.
    packetType                  INTEGER   NOT NULL,                    -- The type of the packet.
    submitSequence              INTEGER   NOT NULL,                    -- Internal sequence starting from 1.
    preempted                   INTEGER   NOT NULL,                    -- True or False if the packet is preempted.
    timeouted                   INTEGER   NOT NULL,                    -- True or False if the packet is timeouted.
    engineType                  INTEGER   NOT NULL                     -- The engine type.
);
CREATE TABLE WDDM_QUEUE_PACKET_INFO_EVENTS (
    -- Raw ETW QueuePacketInfo events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    context                     INTEGER   NOT NULL,                    -- The context ID of WDDM queue.
    packetType                  INTEGER   NOT NULL,                    -- The address of queue packet.
    submitSequence              INTEGER   NOT NULL,                    -- Internal sequence starting from 1.
    engineType                  INTEGER   NOT NULL                     -- The engine type.
);
CREATE TABLE WDDM_DMA_PACKET_START_EVENTS (
    -- Raw ETW DmaPacketStart events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    context                     INTEGER   NOT NULL,                    -- The context ID of WDDM queue.
    queuePacketContext          INTEGER   NOT NULL,                    -- The queue packet context.
    uliSubmissionId             INTEGER   NOT NULL,                    -- The queue packet submission ID.
    dmaBuffer                   INTEGER   NOT NULL,                    -- The reported address of dma buffer.
    packetType                  INTEGER   NOT NULL,                    -- The type of the packet.
    ulQueueSubmitSequence       INTEGER   NOT NULL,                    -- Internal sequence starting from 1.
    quantumStatus               INTEGER   NOT NULL,                    -- The quantum Status.
    engineType                  INTEGER   NOT NULL                     -- The engine type.
);
CREATE TABLE WDDM_DMA_PACKET_STOP_EVENTS (
    -- Raw ETW DmaPacketStop events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    context                     INTEGER   NOT NULL,                    -- The context ID of WDDM queue.
    uliCompletionId             INTEGER   NOT NULL,                    -- The queue packet completion ID.
    packetType                  INTEGER   NOT NULL,                    -- The type of the packet.
    ulQueueSubmitSequence       INTEGER   NOT NULL,                    -- Internal sequence starting from 1.
    preempted                   INTEGER   NOT NULL,                    -- True or False if the packet is preempted.
    engineType                  INTEGER   NOT NULL                     -- The engine type.
);
CREATE TABLE WDDM_DMA_PACKET_INFO_EVENTS (
    -- Raw ETW DmaPacketInfo events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    context                     INTEGER   NOT NULL,                    -- The context ID of WDDM queue.
    uliCompletionId             INTEGER   NOT NULL,                    -- The queue packet completion ID.
    faultedVirtualAddress       INTEGER   NOT NULL,                    -- The virtual address of faulted process.
    faultedProcessHandle        INTEGER   NOT NULL,                    -- The address of faulted process.
    packetType                  INTEGER   NOT NULL,                    -- The type of the packet.
    ulQueueSubmitSequence       INTEGER   NOT NULL,                    -- Internal sequence starting from 1.
    interruptType               INTEGER   NOT NULL,                    -- The DMA interrupt type.
    quantumStatus               INTEGER   NOT NULL,                    -- The quantum Status.
    pageFaultFlags              INTEGER   NOT NULL,                    -- The page fault flag ID.
    engineType                  INTEGER   NOT NULL                     -- The engine type.
);
CREATE TABLE WDDM_HW_QUEUE_EVENTS (
    -- Raw ETW HwQueueStart events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    context                     INTEGER   NOT NULL,                    -- The context ID of WDDM queue.
    hwQueue                     INTEGER   NOT NULL,                    -- The address of HW queue.
    parentDxgHwQueue            INTEGER   NOT NULL                     -- The address of parent Dxg HW queue.
);
CREATE TABLE NVVIDEO_ENCODER_API (
    -- NV Video Encoder API traced to gather information about NVIDIA Video Codek SDK Encoder APIs.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- NVIDIA Video event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id) -- StringId of the function name.
);
CREATE TABLE NVVIDEO_DECODER_API (
    -- NV Video Encoder API traced to gather information about NVIDIA Video Codek SDK Decoder APIs.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    eventClass                  INTEGER   NOT NULL,                    -- NVIDIA Video event class enum value. See docs for specifics.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id) -- StringId of the function name.
);
CREATE TABLE GPU_MEMORY_BUDGET_EVENTS (
    -- Raw ETW VidMmProcessBudgetChange events.

    timestamp                   INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    newBudget                   INTEGER,                               -- The new budget size in bytes.
    segmentGroup                INTEGER                                -- The segment group ID.
);
CREATE TABLE GPU_MEMORY_USAGE_EVENTS (
    -- Raw ETW VidMmProcessUsageChange events.

    timestamp                   INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    newUsage                    INTEGER,                               -- The new usage size in bytes.
    segmentGroup                INTEGER                                -- The segment group ID.
);
CREATE TABLE DEMOTED_BYTES_EVENTS (
    -- Raw ETW VidMmProcessDemotedCommitmentChange events.

    timestamp                   INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    commitment                  INTEGER                                -- Total demoted bytes.
);
CREATE TABLE TOTAL_BYTES_RESIDENT_IN_SEGMENT_EVENTS (
    -- Raw ETW TotalBytesResidentInSegment events.

    timestamp                   INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    totalBytesResident          INTEGER,                               -- Total bytes resident in segment.
    segmentGroup                INTEGER,                               -- The segment group ID.
    segmentId                   INTEGER                                -- The segment ID.
);
CREATE TABLE NV_LOAD_BALANCE_MASTER_EVENTS (
    -- Raw ETW NV-wgf2um LoadBalanceMaster events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    eventId                     INTEGER   NOT NULL,                    -- Event ID.
    task                        TEXT      NOT NULL,                    -- The task name.
    frameCount                  INTEGER   NOT NULL,                    -- The frame ID.
    frameTime                   REAL      NOT NULL,                    -- Frame duration.
    averageFrameTime            REAL      NOT NULL,                    -- Average of frame duration.
    averageLatency              REAL      NOT NULL,                    -- Average of latency.
    minLatency                  REAL      NOT NULL,                    -- The minimum latency.
    averageQueuedFrames         REAL      NOT NULL,                    -- Average number of queued frames.
    totalActiveMs               REAL      NOT NULL,                    -- Total active time in milliseconds.
    totalIdleMs                 REAL      NOT NULL,                    -- Total idle time in milliseconds.
    idlePercent                 REAL      NOT NULL,                    -- The percentage of idle time.
    isGPUAlmostOneFrameAhead    INTEGER   NOT NULL                     -- True or False if GPU is almost one frame ahead.
);
CREATE TABLE NV_LOAD_BALANCE_EVENTS (
    -- Raw ETW NV-wgf2um LoadBalance events.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    globalTid                   INTEGER   NOT NULL,                    -- Serialized GlobalId.
    gpu                         INTEGER   NOT NULL,                    -- GPU index.
    eventId                     INTEGER   NOT NULL,                    -- Event ID.
    task                        TEXT      NOT NULL,                    -- The task name.
    averageFPS                  REAL      NOT NULL,                    -- Average frame per second.
    queuedFrames                REAL      NOT NULL,                    -- The amount of queued frames.
    averageQueuedFrames         REAL      NOT NULL,                    -- Average number of queued frames.
    currentCPUTime              REAL      NOT NULL,                    -- The current CPU time.
    averageCPUTime              REAL      NOT NULL,                    -- Average CPU time.
    averageStallTime            REAL      NOT NULL,                    -- Average of stall time.
    averageCPUIdleTime          REAL      NOT NULL,                    -- Average CPU idle time.
    isGPUAlmostOneFrameAhead    INTEGER   NOT NULL                     -- True or False if GPU is almost one frame ahead.
);
CREATE TABLE PROCESSES (
    -- Names and identifiers of processes captured in the report.

    globalPid                   INTEGER,                               -- Serialized GlobalId.
    pid                         INTEGER,                               -- The process ID.
    name                        TEXT                                   -- The process name.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_OPENACC_DATA (
    -- OpenACC data events collected using CUPTI.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    eventKind                   INTEGER   NOT NULL,                    -- CUPTI OpenACC event kind. See enum CUpti_OpenAccEventKind.
    DeviceType                  INTEGER   NOT NULL,                    -- Device type. See enum acc_device_t.
    lineNo                      INTEGER   NOT NULL,                    -- Line number of the directive or program construct.
    cuDeviceId                  INTEGER   NOT NULL,                    -- CUDA device ID. Valid only if deviceType is acc_device_nvidia.
    cuContextId                 INTEGER   NOT NULL,                    -- CUDA context ID. Valid only if deviceType is acc_device_nvidia.
    cuStreamId                  INTEGER   NOT NULL,                    -- CUDA stream ID. Valid only if deviceType is acc_device_nvidia.
    srcFile                     INTEGER   REFERENCES StringIds(id),    -- StringId of the source file name or path.
    funcName                    INTEGER   REFERENCES StringIds(id),    -- StringId of the function in which the event occurred.
    bytes                       INTEGER,                               -- Number of bytes.
    varName                     INTEGER   REFERENCES StringIds(id)     -- StringId of the variable name.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_OPENACC_LAUNCH (
    -- OpenACC launch events collected using CUPTI.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    eventKind                   INTEGER   NOT NULL,                    -- CUPTI OpenACC event kind. See enum CUpti_OpenAccEventKind.
    DeviceType                  INTEGER   NOT NULL,                    -- Device type. See enum acc_device_t.
    lineNo                      INTEGER   NOT NULL,                    -- Line number of the directive or program construct.
    cuDeviceId                  INTEGER   NOT NULL,                    -- CUDA device ID. Valid only if deviceType is acc_device_nvidia.
    cuContextId                 INTEGER   NOT NULL,                    -- CUDA context ID. Valid only if deviceType is acc_device_nvidia.
    cuStreamId                  INTEGER   NOT NULL,                    -- CUDA stream ID. Valid only if deviceType is acc_device_nvidia.
    srcFile                     INTEGER   REFERENCES StringIds(id),    -- StringId of the source file name or path.
    funcName                    INTEGER   REFERENCES StringIds(id),    -- StringId of the function in which the event occurred.
    numGangs                    INTEGER,                               -- Number of gangs created for this kernel launch.
    numWorkers                  INTEGER,                               -- Number of workers created for this kernel launch.
    vectorLength                INTEGER,                               -- Number of vector lanes created for this kernel launch.
    kernelName                  INTEGER   REFERENCES StringIds(id)     -- StringId of the kernel name.
);
CREATE TABLE CUPTI_ACTIVITY_KIND_OPENACC_OTHER (
    -- OpenACC other events collected using CUPTI.

    start                       INTEGER   NOT NULL,                    -- Event start timestamp (ns).
    end                         INTEGER   NOT NULL,                    -- Event end timestamp (ns).
    nameId                      INTEGER   NOT NULL   REFERENCES StringIds(id), -- StringId of the event name.
    globalTid                   INTEGER,                               -- Serialized GlobalId.
    eventKind                   INTEGER   NOT NULL,                    -- CUPTI OpenACC event kind. See enum CUpti_OpenAccEventKind.
    DeviceType                  INTEGER   NOT NULL,                    -- Device type. See enum acc_device_t.
    lineNo                      INTEGER   NOT NULL,                    -- Line number of the directive or program construct.
    cuDeviceId                  INTEGER   NOT NULL,                    -- CUDA device ID. Valid only if deviceType is acc_device_nvidia.
    cuContextId                 INTEGER   NOT NULL,                    -- CUDA context ID. Valid only if deviceType is acc_device_nvidia.
    cuStreamId                  INTEGER   NOT NULL,                    -- CUDA stream ID. Valid only if deviceType is acc_device_nvidia.
    srcFile                     INTEGER   REFERENCES StringIds(id),    -- StringId of the source file name or path.
    funcName                    INTEGER   REFERENCES StringIds(id)     -- StringId of the function in which the event occurred.
);

Here are the set values stored in enums in the Nsight Systems SQLite schema

CUDA Event Class Values

0 - TRACE_PROCESS_EVENT_CUDA_RUNTIME
1 - TRACE_PROCESS_EVENT_CUDA_DRIVER
13 - TRACE_PROCESS_EVENT_CUDA_EGL_DRIVER
28 - TRACE_PROCESS_EVENT_CUDNN
29 - TRACE_PROCESS_EVENT_CUBLAS
33 - TRACE_PROCESS_EVENT_CUDNN_START
34 - TRACE_PROCESS_EVENT_CUDNN_FINISH
35 - TRACE_PROCESS_EVENT_CUBLAS_START
36 - TRACE_PROCESS_EVENT_CUBLAS_FINISH
67 - TRACE_PROCESS_EVENT_CUDABACKTRACE
77 - TRACE_PROCESS_EVENT_CUDA_GRAPH_NODE_CREATION    
    

See CUPTI documentation for detailed information on collected event and data types.

NVTX Event Type Values

33 - NvtxCategory
34 - NvtxMark
39 - NvtxThread
59 - NvtxPushPopRange
60 - NvtxStartEndRange
75 - NvtxDomainCreate
76 - NvtxDomainDestroy
    

The difference between text and textId columns is that if an NVTX event message was passed via call to nvtxDomainRegisterString function, then the message will be available through textId field, otherwise the text field will contain the message if it was provided.

OpenGL Events

KHR event class values

62 - KhrDebugPushPopRange
63 - KhrDebugGpuPushPopRange
    

KHR source kind values

0x8249 - GL_DEBUG_SOURCE_THIRD_PARTY
0x824A - GL_DEBUG_SOURCE_APPLICATION
    

KHR type values

0x824C - GL_DEBUG_TYPE_ERROR
0x824D - GL_DEBUG_TYPE_DEPRECATED_BEHAVIOR
0x824E - GL_DEBUG_TYPE_UNDEFINED_BEHAVIOR
0x824F - GL_DEBUG_TYPE_PORTABILITY
0x8250 - GL_DEBUG_TYPE_PERFORMANCE
0x8251 - GL_DEBUG_TYPE_OTHER
0x8268 - GL_DEBUG_TYPE_MARKER
0x8269 - GL_DEBUG_TYPE_PUSH_GROUP
0x826A - GL_DEBUG_TYPE_POP_GROUP
   

KHR severity values

0x826B - GL_DEBUG_SEVERITY_NOTIFICATION
0x9146 - GL_DEBUG_SEVERITY_HIGH
0x9147 - GL_DEBUG_SEVERITY_MEDIUM
0x9148 - GL_DEBUG_SEVERITY_LOW
   

OSRT Event Class Values

OS runtime libraries can be traced to gather information about low-level userspace APIs. This traces the system call wrappers and thread synchronization interfaces exposed by the C runtime and POSIX Threads (pthread) libraries. This does not perform a complete runtime library API trace, but instead focuses on the functions that can take a long time to execute, or could potentially cause your thread be unscheduled from the CPU while waiting for an event to complete.

OSRT events may have callchains attached to them, depending on selected profiling settings. In such cases, one can use callchainId column to select relevant callchains from OSRT_CALLCHAINS table

OSRT event class values

27 - TRACE_PROCESS_EVENT_OS_RUNTIME
31 - TRACE_PROCESS_EVENT_OS_RUNTIME_START
32 - TRACE_PROCESS_EVENT_OS_RUNTIME_FINISH
   

DX12 Event Class Values

41 - TRACE_PROCESS_EVENT_DX12_API
42 - TRACE_PROCESS_EVENT_DX12_WORKLOAD
43 - TRACE_PROCESS_EVENT_DX12_START
44 - TRACE_PROCESS_EVENT_DX12_FINISH
52 - TRACE_PROCESS_EVENT_DX12_DISPLAY
59 - TRACE_PROCESS_EVENT_DX12_CREATE_OBJECT
   

PIX Event Class Values

65 - TRACE_PROCESS_EVENT_DX12_DEBUG_API
75 - TRACE_PROCESS_EVENT_DX11_DEBUG_API
   

Vulkan Event Class Values

53 - TRACE_PROCESS_EVENT_VULKAN_API
54 - TRACE_PROCESS_EVENT_VULKAN_WORKLOAD
55 - TRACE_PROCESS_EVENT_VULKAN_START
56 - TRACE_PROCESS_EVENT_VULKAN_FINISH
60 - TRACE_PROCESS_EVENT_VULKAN_CREATE_OBJECT
66 - TRACE_PROCESS_EVENT_VULKAN_DEBUG_API
   

Vulkan Flags

VALID_BIT = 0x00000001
CACHE_HIT_BIT = 0x00000002
BASE_PIPELINE_ACCELERATION_BIT = 0x00000004
   

SLI Event Class Values

62 - TRACE_PROCESS_EVENT_SLI
63 - TRACE_PROCESS_EVENT_SLI_START
64 - TRACE_PROCESS_EVENT_SLI_FINISH
   

SLI Transfer Info Values

0 - P2P_SKIPPED
1 - P2P_EARLY_PUSH
2 - P2P_PUSH_FAILED
3 - P2P_2WAY_OR_PULL
4 - P2P_PRESENT
5 - P2P_DX12_INIT_PUSH_ON_WRITE
    

WDDM Event Values

VIDMM operation type values

0 - None
101 - RestoreSegments
102 - PurgeSegments
103 - CleanupPrimary
104 - AllocatePagingBufferResources
105 - FreePagingBufferResources
106 - ReportVidMmState
107 - RunApertureCoherencyTest
108 - RunUnmapToDummyPageTest
109 - DeferredCommand
110 - SuspendMemorySegmentAccess
111 - ResumeMemorySegmentAccess
112 - EvictAndFlush
113 - CommitVirtualAddressRange
114 - UncommitVirtualAddressRange
115 - DestroyVirtualAddressAllocator
116 - PageInDevice
117 - MapContextAllocation
118 - InitPagingProcessVaSpace
200 - CloseAllocation
202 - ComplexLock
203 - PinAllocation
204 - FlushPendingGpuAccess
205 - UnpinAllocation
206 - MakeResident
207 - Evict
208 - LockInAperture
209 - InitContextAllocation
210 - ReclaimAllocation
211 - DiscardAllocation
212 - SetAllocationPriority
1000 - EvictSystemMemoryOfferList
   

Paging queue type values

0 - VIDMM_PAGING_QUEUE_TYPE_UMD
1 - VIDMM_PAGING_QUEUE_TYPE_Default
2 - VIDMM_PAGING_QUEUE_TYPE_Evict
3 - VIDMM_PAGING_QUEUE_TYPE_Reclaim
   

Packet type values

0 - DXGKETW_RENDER_COMMAND_BUFFER
1 - DXGKETW_DEFERRED_COMMAND_BUFFER
2 - DXGKETW_SYSTEM_COMMAND_BUFFER
3 - DXGKETW_MMIOFLIP_COMMAND_BUFFER
4 - DXGKETW_WAIT_COMMAND_BUFFER
5 - DXGKETW_SIGNAL_COMMAND_BUFFER
6 - DXGKETW_DEVICE_COMMAND_BUFFER
7 - DXGKETW_SOFTWARE_COMMAND_BUFFER
   

Engine type values

0 - DXGK_ENGINE_TYPE_OTHER
1 - DXGK_ENGINE_TYPE_3D
2 - DXGK_ENGINE_TYPE_VIDEO_DECODE
3 - DXGK_ENGINE_TYPE_VIDEO_ENCODE
4 - DXGK_ENGINE_TYPE_VIDEO_PROCESSING
5 - DXGK_ENGINE_TYPE_SCENE_ASSEMBLY
6 - DXGK_ENGINE_TYPE_COPY
7 - DXGK_ENGINE_TYPE_OVERLAY
8 - DXGK_ENGINE_TYPE_CRYPTO
   

DMA interrupt type values

1 = DXGK_INTERRUPT_DMA_COMPLETED
2 = DXGK_INTERRUPT_DMA_PREEMPTED
4 = DXGK_INTERRUPT_DMA_FAULTED
9 = DXGK_INTERRUPT_DMA_PAGE_FAULTED
   

Queue type values

0 = Queue_Packet
1 = Dma_Packet
2 = Paging_Queue_Packet
   

Driver Events

Load balance event type values

1 - LoadBalanceEvent_GPU
8 - LoadBalanceEvent_CPU
21 - LoadBalanceMasterEvent_GPU
22 - LoadBalanceMasterEvent_CPU
   

OpenMP Events

OpenMP event class values

78 - TRACE_PROCESS_EVENT_OPENMP
79 - TRACE_PROCESS_EVENT_OPENMP_START
80 - TRACE_PROCESS_EVENT_OPENMP_FINISH
   

OpenMP event kind values

15 - OPENMP_EVENT_KIND_TASK_CREATE
16 - OPENMP_EVENT_KIND_TASK_SCHEDULE
17 - OPENMP_EVENT_KIND_CANCEL
20 - OPENMP_EVENT_KIND_MUTEX_RELEASED
21 - OPENMP_EVENT_KIND_LOCK_INIT
22 - OPENMP_EVENT_KIND_LOCK_DESTROY
25 - OPENMP_EVENT_KIND_DISPATCH
26 - OPENMP_EVENT_KIND_FLUSH
27 - OPENMP_EVENT_KIND_THREAD
28 - OPENMP_EVENT_KIND_PARALLEL
29 - OPENMP_EVENT_KIND_SYNC_REGION_WAIT
30 - OPENMP_EVENT_KIND_SYNC_REGION
31 - OPENMP_EVENT_KIND_TASK
32 - OPENMP_EVENT_KIND_MASTER
33 - OPENMP_EVENT_KIND_REDUCTION
34 - OPENMP_EVENT_KIND_MUTEX_WAIT
35 - OPENMP_EVENT_KIND_CRITICAL_SECTION
36 - OPENMP_EVENT_KIND_WORKSHARE
   

OpenMP thread type values

1 - OpenMP Initial Thread
2 - OpenMP Worker Thread
3 - OpenMP Internal Thread
4 - Unknown
   

OpenMP sync region kind values

1 - Barrier
2 - Implicit barrier
3 - Explicit barrier
4 - Implementation-dependent barrier
5 - Taskwait
6 - Taskgroup
    

OpenMP task kind values

1 - Initial task
2 - Implicit task
3 - Explicit task
    

OpenMP prior task status values

1 - Task completed
2 - Task yielded to another task
3 - Task was cancelled
7 - Task was switched out for other reasons
    

OpenMP mutex kind values

1 - Waiting for lock
2 - Testing lock
3 - Waiting for nested lock
4 - Tesing nested lock
5 - Waitng for entering critical section region
6 - Waiting for entering atomic region
7 - Waiting for entering ordered region
    

OpenMP critical section kind values

5 - Critical section region
6 - Atomic region
7 - Ordered region
    

OpenMP workshare kind values

1 - Loop region
2 - Sections region
3 - Single region (executor)
4 - Single region (waiting)
5 - Workshare region
6 - Distrubute region
7 - Taskloop region
    

OpenMP dispatch kind values

1 - Iteration
2 - Section
    

Common Helper Commands

When utilizing sqlite3 command line tool, it’s helpful to have data printed as named columns, this can be done with:

.mode column
.headers on
    

Default column width is determined by the data in the first row of results. If this doesn’t work out well, you can specify widths manually.

.width 10 20 50

Obtaining Sample Report

CLI interface of Nsight Systems was used to profile radixSortThrust CUDA sample, then the resulting .nsys-rep file was exported using the nsys export.

nsys profile --trace=cuda,osrt radixSortThrust
nsys export --type sqlite report1.nsys-rep
    

Serialized Process and Thread Identifiers

Nsight Systems stores identifiers where events originated in serialized form. For events that have globalTid or globalPid fields exported, use the following code to extract numeric TID and PID.

SELECT globalTid / 0x1000000 % 0x1000000 AS PID, globalTid % 0x1000000 AS TID FROM TABLE_NAME;
    

Note: globalTid field includes both TID and PID values, while globalPid only containes the PID value.

Correlate CUDA Kernel Launches With CUDA API Kernel Launches

ALTER TABLE CUPTI_ACTIVITY_KIND_RUNTIME ADD COLUMN name TEXT;
ALTER TABLE CUPTI_ACTIVITY_KIND_RUNTIME ADD COLUMN kernelName TEXT;

UPDATE CUPTI_ACTIVITY_KIND_RUNTIME SET kernelName =
    (SELECT value FROM StringIds
    JOIN CUPTI_ACTIVITY_KIND_KERNEL AS cuda_gpu
        ON cuda_gpu.shortName = StringIds.id
        AND CUPTI_ACTIVITY_KIND_RUNTIME.correlationId = cuda_gpu.correlationId);

UPDATE CUPTI_ACTIVITY_KIND_RUNTIME SET name =
    (SELECT value FROM StringIds WHERE nameId = StringIds.id);
    

Select 10 longest CUDA API ranges that resulted in kernel execution.

    
SELECT name, kernelName, start, end FROM CUPTI_ACTIVITY_KIND_RUNTIME
    WHERE kernelName IS NOT NULL ORDER BY end - start LIMIT 10;
    

Results:

name                    kernelName               start       end       
----------------------  -----------------------  ----------  ----------
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  658863435   658868490 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  609755015   609760075 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  632683286   632688349 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  606495356   606500439 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  603114486   603119586 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  802729785   802734906 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  593381170   593386294 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  658759955   658765090 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  681549917   681555059 
cudaLaunchKernel_v7000  RadixSortScanBinsKernel  717812527   717817671
    

Remove Ranges Overlapping With Overhead

Use the this query to count CUDA API ranges overlapping with the overhead ones.

Replace "SELECT COUNT(*)" with "DELETE" to remove such ranges.

SELECT COUNT(*) FROM CUPTI_ACTIVITY_KIND_RUNTIME WHERE rowid IN
(
    SELECT cuda.rowid
    FROM PROFILER_OVERHEAD as overhead
    INNER JOIN CUPTI_ACTIVITY_KIND_RUNTIME as cuda ON
    (cuda.start BETWEEN overhead.start and overhead.end)
    OR (cuda.end BETWEEN overhead.start and overhead.end)
    OR (cuda.start < overhead.start AND cuda.end > overhead.end)
);
    

Results:

COUNT(*)  
----------
1095      
    

Find CUDA API Calls That Resulted in Original Graph Node Creation.

SELECT graph.graphNodeId, api.start, graph.start as graphStart, api.end,
    api.globalTid, api.correlationId, api.globalTid,
    (SELECT value FROM StringIds where api.nameId == id) as name
FROM CUPTI_ACTIVITY_KIND_RUNTIME as api
JOIN
    (
        SELECT start, graphNodeId, globalTid from CUDA_GRAPH_EVENTS
        GROUP BY graphNodeId
        HAVING COUNT(originalGraphNodeId) = 0
    ) as graph
ON api.globalTid == graph.globalTid AND api.start < graph.start AND api.end > graph.start
ORDER BY graphNodeId;
    

Results:

graphNodeId  start       graphStart  end         globalTid        correlationId  globalTid        name                         
-----------  ----------  ----------  ----------  ---------------  -------------  ---------------  -----------------------------
1            584366518   584378040   584379102   281560221750233  109            281560221750233  cudaGraphAddMemcpyNode_v10000
2            584379402   584382428   584383139   281560221750233  110            281560221750233  cudaGraphAddMemsetNode_v10000
3            584390663   584395352   584396053   281560221750233  111            281560221750233  cudaGraphAddKernelNode_v10000
4            584396314   584397857   584398438   281560221750233  112            281560221750233  cudaGraphAddMemsetNode_v10000
5            584398759   584400311   584400812   281560221750233  113            281560221750233  cudaGraphAddKernelNode_v10000
6            584401083   584403047   584403527   281560221750233  114            281560221750233  cudaGraphAddMemcpyNode_v10000
7            584403928   584404920   584405491   281560221750233  115            281560221750233  cudaGraphAddHostNode_v10000  
29           632107852   632117921   632121407   281560221750233  144            281560221750233  cudaMemcpyAsync_v3020        
30           632122168   632125545   632127989   281560221750233  145            281560221750233  cudaMemsetAsync_v3020        
31           632131546   632133339   632135584   281560221750233  147            281560221750233  cudaMemsetAsync_v3020        
34           632162514   632167393   632169297   281560221750233  151            281560221750233  cudaMemcpyAsync_v3020        
35           632170068   632173334   632175388   281560221750233  152            281560221750233  cudaLaunchHostFunc_v10000    
   

Backtraces for OSRT Ranges

Adding text columns makes results of the query below more human-readable.

ALTER TABLE OSRT_API ADD COLUMN name TEXT;
UPDATE OSRT_API SET name = (SELECT value FROM StringIds WHERE OSRT_API.nameId = StringIds.id);

ALTER TABLE OSRT_CALLCHAINS ADD COLUMN symbolName TEXT;
UPDATE OSRT_CALLCHAINS SET symbolName = (SELECT value FROM StringIds WHERE symbol = StringIds.id);

ALTER TABLE OSRT_CALLCHAINS ADD COLUMN moduleName TEXT;
UPDATE OSRT_CALLCHAINS SET moduleName = (SELECT value FROM StringIds WHERE module = StringIds.id);
   

Print backtrace of the longest OSRT range

SELECT globalTid / 0x1000000 % 0x1000000 AS PID, globalTid % 0x1000000 AS TID,
    start, end, name, callchainId, stackDepth, symbolName, moduleName
FROM OSRT_API LEFT JOIN OSRT_CALLCHAINS ON callchainId == OSRT_CALLCHAINS.id
WHERE OSRT_API.rowid IN (SELECT rowid FROM OSRT_API ORDER BY end - start DESC LIMIT 1)
ORDER BY stackDepth LIMIT 10;
   

Results:

PID         TID         start       end         name                    callchainId  stackDepth  symbolName                      moduleName                              
----------  ----------  ----------  ----------  ----------------------  -----------  ----------  ------------------------------  ----------------------------------------
19163       19176       360897690   860966851   pthread_cond_timedwait  88           0           pthread_cond_timedwait@GLIBC_2  /lib/x86_64-linux-gnu/libpthread-2.27.so
19163       19176       360897690   860966851   pthread_cond_timedwait  88           1           0x7fbc983b7227                  /usr/lib/x86_64-linux-gnu/libcuda.so.418
19163       19176       360897690   860966851   pthread_cond_timedwait  88           2           0x7fbc9835d5c7                  /usr/lib/x86_64-linux-gnu/libcuda.so.418
19163       19176       360897690   860966851   pthread_cond_timedwait  88           3           0x7fbc983b64a8                  /usr/lib/x86_64-linux-gnu/libcuda.so.418
19163       19176       360897690   860966851   pthread_cond_timedwait  88           4           start_thread                    /lib/x86_64-linux-gnu/libpthread-2.27.so
19163       19176       360897690   860966851   pthread_cond_timedwait  88           5           __clone                         /lib/x86_64-linux-gnu/libc-2.27.so      
   

Profiled processes output streams

ALTER TABLE ProcessStreams ADD COLUMN filename TEXT;
UPDATE ProcessStreams SET filename = (SELECT value FROM StringIds WHERE ProcessStreams.filenameId = StringIds.id);

ALTER TABLE ProcessStreams ADD COLUMN content TEXT;
UPDATE ProcessStreams SET content = (SELECT value FROM StringIds WHERE ProcessStreams.contentId = StringIds.id);
   

Select all collected stdout and stderr streams.

select globalPid / 0x1000000 % 0x1000000 AS PID, filename, content from ProcessStreams;
   

Results:

PID         filename                                                 content                                                                                                                                                                                                                                                                                                                    
----------  -------------------------------------------------------  --------------------------------------------------------------------------------------------------------------------                                                                                                                                                                                                       
19163       /tmp/nvidia/nsight_systems/streams/pid_19163_stdout.log  /home/user_name/NVIDIA_CUDA-10.1_Samples/6_Advanced/radixSortThrust/radixSortThrust Starting...

GPU Device 0: "Quadro P2000" with compute capability 6.1


Sorting 1048576 32-bit unsigned int keys and values

radixSortThrust, Throughput = 401.0872 MElements/s, Time = 0.00261 s, Size = 1048576 elements
Test passed

19163       /tmp/nvidia/nsight_systems/streams/pid_19163_stderr.log
   

Thread Summary

Please note, that Nsight Systems applies additional logic during sampling events processing to work around lost events. This means that the results of the below query might differ slightly from the ones shown in “Analysis summary” tab.

Thread summary calculated using CPU cycles (when available).

SELECT
    globalTid / 0x1000000 % 0x1000000 AS PID,
    globalTid % 0x1000000 AS TID,
    ROUND(100.0 * SUM(cpuCycles) / 
        (
            SELECT SUM(cpuCycles) FROM COMPOSITE_EVENTS 
            GROUP BY globalTid / 0x1000000000000 % 0x100
        ),
        2
    ) as CPU_utilization,
    (SELECT value FROM StringIds WHERE id = 
        (
            SELECT nameId FROM ThreadNames 
            WHERE ThreadNames.globalTid = COMPOSITE_EVENTS.globalTid
        )
    ) as thread_name
FROM COMPOSITE_EVENTS
GROUP BY globalTid
ORDER BY CPU_utilization DESC
LIMIT 10;
   

Results:

PID         TID         CPU_utilization  thread_name    
----------  ----------  ---------------  ---------------
19163       19163       98.4             radixSortThrust
19163       19168       1.35             CUPTI worker th
19163       19166       0.25             [NS]    
   

Thread running time may be calculated using scheduling data, when PMU counter data was not collected.

CREATE INDEX sched_start ON SCHED_EVENTS (start);

CREATE TABLE CPU_USAGE AS
SELECT
    first.globalTid as globalTid,
    (SELECT nameId FROM ThreadNames WHERE ThreadNames.globalTid = first.globalTid) as nameId,
    sum(second.start - first.start) as total_duration,
    count() as ranges_count
FROM SCHED_EVENTS as first
LEFT JOIN SCHED_EVENTS as second
ON second.rowid =
    (
        SELECT rowid
        FROM SCHED_EVENTS
        WHERE start > first.start AND globalTid = first.globalTid
        ORDER BY start ASC
        LIMIT 1
    )
WHERE first.isSchedIn != 0
GROUP BY first.globalTid
ORDER BY total_duration DESC;

SELECT
    globalTid / 0x1000000 % 0x1000000 AS PID,
    globalTid % 0x1000000 AS TID,
    (SELECT value FROM StringIds where nameId == id) as thread_name,
    ROUND(100.0 * total_duration / (SELECT SUM(total_duration) FROM CPU_USAGE), 2) as CPU_utilization
FROM CPU_USAGE
ORDER BY CPU_utilization DESC;
   

Results:

PID         TID         thread_name      CPU_utilization
----------  ----------  ---------------  ---------------
19163       19163       radixSortThrust  93.74          
19163       19169       radixSortThrust  3.22           
19163       19168       CUPTI worker th  2.46           
19163       19166       [NS]             0.44           
19163       19172       radixSortThrust  0.07           
19163       19167       [NS Comms]       0.05           
19163       19176       radixSortThrust  0.02           
19163       19170       radixSortThrust  0.0
   

Function Table

These examples demonstrate how to calculate Flat and BottomUp (for top level only) views statistics.

To set up:

ALTER TABLE SAMPLING_CALLCHAINS ADD COLUMN symbolName TEXT;
UPDATE SAMPLING_CALLCHAINS SET symbolName = (SELECT value FROM StringIds WHERE symbol = StringIds.id);

ALTER TABLE SAMPLING_CALLCHAINS ADD COLUMN moduleName TEXT;
UPDATE SAMPLING_CALLCHAINS SET moduleName = (SELECT value FROM StringIds WHERE module = StringIds.id);
   

To get flat view:

SELECT symbolName, moduleName, ROUND(100.0 * sum(cpuCycles) /
    (SELECT SUM(cpuCycles) FROM COMPOSITE_EVENTS), 2) AS flatTimePercentage
FROM SAMPLING_CALLCHAINS
LEFT JOIN COMPOSITE_EVENTS ON SAMPLING_CALLCHAINS.id == COMPOSITE_EVENTS.id
GROUP BY symbol, module
ORDER BY flatTimePercentage DESC
LIMIT 5;
   

To get BottomUp view (top level only):

SELECT symbolName, moduleName, ROUND(100.0 * sum(cpuCycles) /
    (SELECT SUM(cpuCycles) FROM COMPOSITE_EVENTS), 2) AS selfTimePercentage
FROM SAMPLING_CALLCHAINS
LEFT JOIN COMPOSITE_EVENTS ON SAMPLING_CALLCHAINS.id == COMPOSITE_EVENTS.id
WHERE stackDepth == 0
GROUP BY symbol, module
ORDER BY selfTimePercentage DESC
LIMIT 5;
   

Results:

symbolName   moduleName   flatTimePercentage
-----------  -----------  ------------------
[Max depth]  [Max depth]  99.92             
thrust::zip  /home/user_  24.17             
thrust::zip  /home/user_  24.17             
thrust::det  /home/user_  24.17             
thrust::det  /home/user_  24.17             
symbolName      moduleName                                   selfTimePercentage
--------------  -------------------------------------------  ------------------
0x7fbc984982b6  /usr/lib/x86_64-linux-gnu/libcuda.so.418.39  5.29              
0x7fbc982d0010  /usr/lib/x86_64-linux-gnu/libcuda.so.418.39  2.81              
thrust::iterat  /home/user_name/NVIDIA_CUDA-10.1_Samples/6_  2.23              
thrust::iterat  /home/user_name/NVIDIA_CUDA-10.1_Samples/6_  1.55              
void thrust::i  /home/user_name/NVIDIA_CUDA-10.1_Samples/6_  1.55
   

DX12 API Frame Duration Histogram

The example demonstrates how to calculate DX12 CPU frames durartion and construct a histogram out of it.

CREATE INDEX DX12_API_ENDTS ON DX12_API (end);

CREATE TEMP VIEW DX12_API_FPS AS SELECT end AS start,
    (SELECT end FROM DX12_API
        WHERE end > outer.end AND nameId == (SELECT id FROM StringIds
            WHERE value == "IDXGISwapChain::Present")
        ORDER BY end ASC LIMIT 1) AS end
FROM DX12_API AS outer
    WHERE nameId == (SELECT id FROM StringIds WHERE value == "IDXGISwapChain::Present")
ORDER BY end;
   

Number of frames with a duration of [X, X + 1) milliseconds.

SELECT
    CAST((end - start) / 1000000.0 AS INT) AS duration_ms,
    count(*)
FROM DX12_API_FPS
WHERE end IS NOT NULL
GROUP BY duration_ms
ORDER BY duration_ms;
   

Results:

duration_ms  count(*)  
-----------  ----------
3            1         
4            2         
5            7         
6            153       
7            19        
8            116       
9            16        
10           8         
11           2         
12           2         
13           1         
14           4         
16           3         
17           2         
18           1
   

GPU Context Switch Events Enumeration

GPU context duration is between first BEGIN and a matching END event.

SELECT (CASE tag WHEN 8 THEN "BEGIN" WHEN 7 THEN "END" END) AS tag,
    globalPid / 0x1000000 % 0x1000000 AS PID,
    vmId, seqNo, contextId, timestamp, gpuId FROM FECS_EVENTS
WHERE tag in (7, 8) ORDER BY seqNo LIMIT 10;
   

Results:

tag         PID         vmId        seqNo       contextId   timestamp   gpuId     
----------  ----------  ----------  ----------  ----------  ----------  ----------
BEGIN       23371       0           0           1048578     56759171    0         
BEGIN       23371       0           1           1048578     56927765    0         
BEGIN       23371       0           3           1048578     63799379    0         
END         23371       0           4           1048578     63918806    0         
BEGIN       19397       0           5           1048577     64014692    0         
BEGIN       19397       0           6           1048577     64250369    0         
BEGIN       19397       0           8           1048577     1918310004  0         
END         19397       0           9           1048577     1918521098  0         
BEGIN       19397       0           10          1048577     2024164744  0         
BEGIN       19397       0           11          1048577     2024358650  0   
   

Resolve NVTX Category Name

The example demonstrates how to resolve NVTX category name for NVTX marks and ranges.

WITH
  event AS (
    SELECT *
    FROM NVTX_EVENTS
    WHERE eventType IN (34, 59, 60) -- mark, push/pop, start/end
  ),
  category AS (
    SELECT
      category,
      domainId,
      text AS categoryName
    FROM NVTX_EVENTS
    WHERE eventType == 33 --  new category
  )
SELECT
  start,
  end,
  globalTid,
  eventType,
  domainId,
  category,
  categoryName,
  text
FROM event JOIN category USING (category, domainId)
ORDER BY start;
   

Results:

start       end         globalTid        eventType   domainId    category    categoryName               text            
----------  ----------  ---------------  ----------  ----------  ----------  -------------------------  ----------------
18281150    18311960    281534938484214  59          0           1           FirstCategoryUnderDefault  Push Pop Range A
18288187    18306674    281534938484214  59          0           2           SecondCategoryUnderDefaul  Push Pop Range B
18294247                281534938484214  34          0           1           FirstCategoryUnderDefault  Mark A          
18300034                281534938484214  34          0           2           SecondCategoryUnderDefaul  Mark B          
18345546    18372595    281534938484214  60          1           1           FirstCategoryUnderMyDomai  Start End Range 
18352924    18378342    281534938484214  60          1           2           SecondCategoryUnderMyDoma  Start End Range 
18359634                281534938484214  34          1           1           FirstCategoryUnderMyDomai  Mark A          
18365448                281534938484214  34          1           2           SecondCategoryUnderMyDoma  Mark B 
   

Rename CUDA Kernels with NVTX

The example demonstrates how to map innermost NVTX push-pop range to a matching CUDA kernel run.

ALTER TABLE CUPTI_ACTIVITY_KIND_KERNEL ADD COLUMN nvtxRange TEXT;
CREATE INDEX nvtx_start ON NVTX_EVENTS (start);


UPDATE CUPTI_ACTIVITY_KIND_KERNEL SET nvtxRange = (
    SELECT NVTX_EVENTS.text
    FROM NVTX_EVENTS JOIN CUPTI_ACTIVITY_KIND_RUNTIME ON
        NVTX_EVENTS.eventType == 59 AND
        NVTX_EVENTS.globalTid == CUPTI_ACTIVITY_KIND_RUNTIME.globalTid AND
        NVTX_EVENTS.start <= CUPTI_ACTIVITY_KIND_RUNTIME.start AND
        NVTX_EVENTS.end >= CUPTI_ACTIVITY_KIND_RUNTIME.end
    WHERE
        CUPTI_ACTIVITY_KIND_KERNEL.correlationId == CUPTI_ACTIVITY_KIND_RUNTIME.correlationId
    ORDER BY NVTX_EVENTS.start DESC LIMIT 1
);

SELECT start, end, globalPid, StringIds.value as shortName, nvtxRange
FROM CUPTI_ACTIVITY_KIND_KERNEL JOIN StringIds ON shortName == id
ORDER BY start LIMIT 6;
   

Results:

start       end         globalPid          shortName      nvtxRange 
----------  ----------  -----------------  -------------  ----------
526545376   526676256   72057700439031808  MatrixMulCUDA            
526899648   527030368   72057700439031808  MatrixMulCUDA  Add       
527031648   527162272   72057700439031808  MatrixMulCUDA  Add       
527163584   527294176   72057700439031808  MatrixMulCUDA  My Kernel 
527296160   527426592   72057700439031808  MatrixMulCUDA  My Range  
527428096   527558656   72057700439031808  MatrixMulCUDA   
   

Select CUDA Calls With Backtraces

ALTER TABLE CUPTI_ACTIVITY_KIND_RUNTIME ADD COLUMN name TEXT;
UPDATE CUPTI_ACTIVITY_KIND_RUNTIME SET name = (SELECT value FROM StringIds WHERE CUPTI_ACTIVITY_KIND_RUNTIME.nameId = StringIds.id);

ALTER TABLE CUDA_CALLCHAINS ADD COLUMN symbolName TEXT;
UPDATE CUDA_CALLCHAINS SET symbolName = (SELECT value FROM StringIds WHERE symbol = StringIds.id);

SELECT globalTid % 0x1000000 AS TID,
    start, end, name, callchainId, stackDepth, symbolName
FROM CUDA_CALLCHAINS JOIN CUPTI_ACTIVITY_KIND_RUNTIME ON callchainId == CUDA_CALLCHAINS.id
ORDER BY callchainId, stackDepth LIMIT 11;
   

Results:

TID         start       end         name           callchainId  stackDepth  symbolName    
----------  ----------  ----------  -------------  -----------  ----------  --------------
11928       168976467   169077826   cuMemAlloc_v2  1            0           0x7f13c44f02ab
11928       168976467   169077826   cuMemAlloc_v2  1            1           0x7f13c44f0b8f
11928       168976467   169077826   cuMemAlloc_v2  1            2           0x7f13c44f3719
11928       168976467   169077826   cuMemAlloc_v2  1            3           cuMemAlloc_v2 
11928       168976467   169077826   cuMemAlloc_v2  1            4           cudart::driver
11928       168976467   169077826   cuMemAlloc_v2  1            5           cudart::cudaAp
11928       168976467   169077826   cuMemAlloc_v2  1            6           cudaMalloc    
11928       168976467   169077826   cuMemAlloc_v2  1            7           cudaError cuda
11928       168976467   169077826   cuMemAlloc_v2  1            8           main          
11928       168976467   169077826   cuMemAlloc_v2  1            9           __libc_start_m
11928       168976467   169077826   cuMemAlloc_v2  1            10          _start  
   

SLI Peer-to-Peer Query

The example demonstrates how to query SLI Peer-to-Peer events with resource size greater than value and within a time range sorted by resource size descending.

SELECT *
FROM SLI_P2P
WHERE resourceSize < 98304 AND start > 1568063100 AND end < 1579468901
ORDER BY resourceSize DESC;
   

Results:

start       end         eventClass  globalTid          gpu         frameId     transferSkipped  srcGpu      dstGpu      numSubResources  resourceSize  subResourceIdx  smplWidth   smplHeight  smplDepth   bytesPerElement  dxgiFormat  logSurfaceNames  transferInfo  isEarlyPushManagedByNvApi  useAsyncP2pForResolve  transferFuncName  regimeName  debugName   bindType  
----------  ----------  ----------  -----------------  ----------  ----------  ---------------  ----------  ----------  ---------------  ------------  --------------  ----------  ----------  ----------  ---------------  ----------  ---------------  ------------  -------------------------  ---------------------  ----------------  ----------  ----------  ----------
1570351100  1570351101  62          72057698056667136  0           771         0                256         512         1                1048576       0               256         256         1           16               2                            3             0                          0                                                                          
1570379300  1570379301  62          72057698056667136  0           771         0                256         512         1                1048576       0               64          64          64          4                31                           3             0                          0                                                                          
1572316400  1572316401  62          72057698056667136  0           773         0                256         512         1                1048576       0               256         256         1           16               2                            3             0                          0                                                                          
1572345400  1572345401  62          72057698056667136  0           773         0                256         512         1                1048576       0               64          64          64          4                31                           3             0                          0                                                                          
1574734300  1574734301  62          72057698056667136  0           775         0                256         512         1                1048576       0               256         256         1           16               2                            3             0                          0                                                                          
1574767200  1574767201  62          72057698056667136  0           775         0                256         512         1                1048576       0               64          64          64          4                31                           3             0                          0                                                                          
   

Generic Events

Syscall usage histogram by PID:

SELECT json_extract(data, '$.common_pid') AS PID, count(*) AS total
FROM GENERIC_EVENTS WHERE PID IS NOT NULL AND typeId = (
  SELECT typeId FROM GENERIC_EVENT_TYPES
  WHERE json_extract(data, '$.Name') = "raw_syscalls:sys_enter")
GROUP BY PID
ORDER BY total DESC
LIMIT 10;
   

Results:

PID         total     
----------  ----------
5551        32811     
9680        3988      
4328        1477      
9564        1246      
4376        1204      
4377        1167      
4357        656       
4355        655       
4356        640       
4354        633
   

Fetching Generic Events in JSON Format

Text and JSON export modes don’t include generic events. Use the below queries (without LIMIT clause) to extract JSON lines representation of generic events, types and sources.

SELECT json_insert('{}',
    '$.sourceId', sourceId,
    '$.data', json(data)
)
FROM GENERIC_EVENT_SOURCES LIMIT 2;

SELECT json_insert('{}',
    '$.typeId', typeId,
    '$.sourceId', sourceId,
    '$.data', json(data)
)
FROM GENERIC_EVENT_TYPES LIMIT 2;

SELECT json_insert('{}',
    '$.rawTimestamp', rawTimestamp,
    '$.timestamp', timestamp,
    '$.typeId', typeId,
    '$.data', json(data)
)
FROM GENERIC_EVENTS LIMIT 2;
   

Results:

json_insert('{}',
    '$.sourceId', sourceId,
    '$.data', json(data)
)                                       
---------------------------------------------------------------------------------------------------------------
{"sourceId":72057602627862528,"data":{"Name":"FTrace","TimeSource":"ClockMonotonicRaw","SourceGroup":"FTrace"}}
json_insert('{}',
    '$.typeId', typeId,
    '$.sourceId', sourceId,
    '$.data', json(data)
)                                                                                                                                                                                                                                             
--------------------------------------------------------------------------------------------------------------------                                                                                                                                                                                                                                                                                                                                                                                                                                                                                             
{"typeId":72057602627862547,"sourceId":72057602627862528,"data":{"Name":"raw_syscalls:sys_enter","Format":"\"NR %ld (%lx, %lx, %lx, %lx, %lx, %lx)\", REC->id, REC->args[0], REC->args[1], REC->args[2], REC->args[3], REC->args[4], REC->args[5]","Fields":[{"Name":"common_pid","Prefix":"int","Suffix":""},{"Name":"id","Prefix":"long","S
{"typeId":72057602627862670,"sourceId":72057602627862528,"data":{"Name":"irq:irq_handler_entry","Format":"\"irq=%d name=%s\", REC->irq, __get_str(name)","Fields":[{"Name":"common_pid","Prefix":"int","Suffix":""},{"Name":"irq","Prefix":"int","Suffix":""},{"Name":"name","Prefix":"__data_loc char[]","Suffix":""},{"Name":"common_type",
json_insert('{}',
    '$.rawTimestamp', rawTimestamp,
    '$.timestamp', timestamp,
    '$.typeId', typeId,
    '$.data', json(data)
)                                                                 
--------------------------------------------------------------------------------------------------------------------                                                                                   
{"rawTimestamp":1183694330725221,"timestamp":6236683,"typeId":72057602627862670,"data":{"common_pid":"0","irq":"66","name":"327696","common_type":"142","common_flags":"9","common_preempt_count":"0"}}
{"rawTimestamp":1183694333695687,"timestamp":9207149,"typeId":72057602627862670,"data":{"common_pid":"0","irq":"66","name":"327696","common_type":"142","common_flags":"9","common_preempt_count":"0"}}

The Arrow type exported file uses the IPC stream format to store the data in a file. The tables can be read by opening the file as an arrow stream. For example one can use the open_stream function from the arrow python package. For more information on the interfaces that can be used to read an IPC stream file, please refer to the Apache Arrow documentation [1, 2].

The name of each table is included in the schema metadata. Thus, while reading each table, the user can extract the table title from the metadata. The table name metadata field has the key table_name. The titles of all the available tables can be found in section SQLite Schema Reference.

JSON and TXT export formats are generated by serializing buffered messages, each on a new line. First, all collected events are processed. Then strings are serialized, followed by stdout, stderr streams if any, followed by thread names.

Output layout:

{Event #1}
{Event #2}
...
{Event #N}
{Strings}
{Streams}
{Threads}
    

For easier grepping of JSON output, the --separate-strings switch may be used to force manual splitting of strings, streams and thread names data.

Example line split: nsys export --export-json --separate-strings sample.nsys-rep -- -

{"type":"String","id":"3720","value":"Process 14944 was launched by the profiler"}
{"type":"String","id":"3721","value":"Profiling has started."}
{"type":"String","id":"3722","value":"Profiler attached to the process."}
{"type":"String","id":"3723","value":"Profiling has stopped."}
{"type":"ThreadName","globalTid":"72057844756653436","nameId":"14","priority":"10"}
{"type":"ThreadName","globalTid":"72057844756657940","nameId":"15","priority":"10"}
{"type":"ThreadName","globalTid":"72057844756654400","nameId":"24","priority":"10"}
    

Compare with: nsys export --export-json sample.nsys-rep -- -

{"data":["[Unknown]","[Unknown kernel module]","[Max depth]","[Broken backtraces]",
  "[Called from Java]","QnxKernelTrace","mm_","task_submit","class_id","syncpt_id",
  "syncpt_thresh","pid","tid","FTrace","[NSys]","[NSys Comms]", "..." ,"Process
  14944 was launched by the profiler","Profiling has started.","Profiler attached
  to the process.","Profiling has stopped."]}
{"data":[{"nameIdx":"14","priority":"10","globalTid":"72057844756653436"},
  {"nameIdx":"15","priority":"10","globalTid":"72057844756657940"},{"nameIdx":"24",
  "priority":"10","globalTid":"72057844756654400"}]}
    

Note, that only last few lines are shown here for clarity and that carriage returns and indents were added to avoid wrapping documentation.

The Nsight Systems development team created and maintains a set of report scripts for some of the commonly requested reports. These scripts will be updated to adapt to any changes in SQLite schema or internal data structures.

These scripts are located in the Nsight Systems package in the Target-<architecture>/reports directory. The following standard reports are available:

The following formats are available in Nsight Systems

The nvprof command of the Nsight Systems CLI is intended to help former nvprof users transition to nsys. Many nvprof switches are not supported by nsys, often because they are now part of NVIDIA Nsight Compute.

The full nvprof documentation can be found at https://docs.nvidia.com/cuda/profiler-users-guide.

The nvprof transition guide for Nsight Compute can be found at https://docs.nvidia.com/nsight-compute/NsightComputeCli/index.html#nvprof-guide.

Any nvprof switch not listed below is not supported by the nsys nvprof command. No additional nsys functionality is available through this command. New features will not be added to this command in the future.

After choosing the nvprof command switch, the following options are available. When you are ready to move to using Nsight Systems CLI directly, see Command Line Options documentation for the nsys switch(es) given below. Note that the nsys implementation and output may vary from nvprof.

Usage.

nsys nvprof [options]

NVIDIA Visual Profiler (NVVP) and NVIDIA nvprof are deprecated. New GPUs and features will not be supported by those tools. We encourage you to make the move to Nsight Systems now. For additional information, suggestions, and rationale, see the blog series in Other Resources.

Nsight Systems provides a build script to build a self isolated Docker container with the Nsight Systems GUI and VNC server.

You can find the build.py script in the host-linux-x64/Scripts/VncContainer directory (or similar on other architectures) under your Nsight Systems installation directory. You will need to have Docker, and Python 3.5 or later.

Available Parameters

Ports

These ports can be published from the container to provide access to the Docker container:

Volumes

Environment variables

Examples

With VNC access on port 5916:

sudo docker run -p 5916:5900/tcp -ti nsys-ui-vnc:1.0

With VNC access on port 5916 and HTTP access on port 8080:

sudo docker run -p 5916:5900/tcp -p 8080:80/tcp -ti nsys-ui-vnc:1.0

With VNC access on port 5916, HTTP access on port 8080 and RDP access on port 33890:

sudo docker run -p 5916:5900/tcp -p 8080:80/tcp -p 33890:3389/tcp -ti nsys-ui-vnc:1.0

With VNC access on port 5916, shared "HOME" folder from the host, VNC server resolution 3840x2160, and custom VNC password

sudo docker run -p 5916:5900/tcp -v $HOME:/mnt/host/home -e NSYS_WINDOW_WIDTH=3840 -e NSYS_WINDOW_HEIGHT=2160 -e VNC_PASSWORD=7654321 -ti nsys-ui-vnc:1.0

With VNC access on port 5916, shared "HOME" folder from the host, and the projects folder to access reports created by Nsight Systems GUI in container

sudo docker run -p 5916:5900/tcp -v $HOME:/mnt/host/home -v /opt/NsysProjects:/mnt/host/Projects -ti nsys-ui-vnc:1.0

Nsight Systems can capture information about Direct3D 11 API calls made by the profiled process. This includes capturing the execution time of D3D11 API functions, performance markers, and frame durations.

SLI Trace

Trace SLI queries and peer-to-peer transfers of D3D11 applications. Requires SLI hardware and an active SLI profile definition in the NVIDIA console.

Direct3D 12 is a low-overhead 3D graphics and compute API for Microsoft Windows. Information about Direct3D 12 can be found at the Direct3D 12 Programming Guide.

Nsight Systems can capture information about Direct3D 12 usage by the profiled process. This includes capturing the execution time of D3D12 API functions, corresponding workloads executed on the GPU, performance markers, and frame durations.

The Command List Creation row displays time periods when command lists were being created. This enables developers to improve their application’s multi-threaded command list creation. Command list creation time period is measured between the call to ID3D12GraphicsCommandList::Reset and the call to ID3D12GraphicsCommandList::Close.

The GPU row shows a compressed view of the D3D12 queue activity, color-coded by the queue type. Expanding it will show the individual queues and their corresponding API calls.

A Command Queue row is displayed for each D3D12 command queue created by the profiled application. The row’s header displays the queue's running index and its type (Direct, Compute, Copy).

The DX12 API Memory Ops row displays all API memory operations and non-persistent resource mappings. Event ranges in the row are color-coded by the heap type they belong to (Default, Readback, Upload, Custom, or CPU-Visible VRAM), with usage warnings highlighted in yellow. A breakdown of the operations can be found by expanding the row to show rows for each individual heap type.

The following operations and warnings are shown:

• Calls to ID3D12Device::CreateCommittedResource, ID3D12Device4::CreateCommittedResource1, and ID3D12Device8::CreateCommittedResource2

  • A warning will be reported if D3D12_HEAP_FLAG_CREATE_NOT_ZEROED is not set in the method's HeapFlags parameter

• Calls to ID3D12Device::CreateHeap and ID3D12Device4::CreateHeap1

  • A warning will be reported if D3D12_HEAP_FLAG_CREATE_NOT_ZEROED is not set in the Flags field of the method's pDesc parameter

• Calls to ID3D12Resource::ReadFromSubResource

  • A warning will be reported if the read is to a D3D12_CPU_PAGE_PROPERTY_WRITE_COMBINE CPU page or from a D3D12_HEAP_TYPE_UPLOAD resource

• Calls to ID3D12Resource::WriteToSubResource

  • A warning will be reported if the write is from a D3D12_CPU_PAGE_PROPERTY_WRITE_BACK CPU page or to a D3D12_HEAP_TYPE_READBACK resource

• Calls to ID3D12Resource::Map and ID3D12Resource::Unmap will be matched into [Map, Unmap] ranges for non-persistent mappings. If a mapping range is nested, only the most external range (reference count = 1) will be shown.

The API row displays time periods where ID3D12CommandQueue::ExecuteCommandLists was called. The GPU Workload row displays time periods where workloads were executed by the GPU. The workload’s type (Graphics, Compute, Copy, etc.) is displayed on the bar representing the workload’s GPU execution.

In addition, you can see the PIX command queue CPU-side performance markers, GPU-side performance markers and the GPU Command List performance markers, each in their row.

Clicking on a GPU workload highlights the corresponding ID3D12CommandQueue::ExecuteCommandLists, ID3D12GraphicsCommandList::Reset and ID3D12GraphicsCommandList::Close API calls, and vice versa.

Clicking one of these will highlight it and the corresponding other two calls.

Vulkan is a low-overhead, cross-platform 3D graphics and compute API, targeting a wide variety of devices from PCs to mobile phones and embedded platforms. The Vulkan API is defined by the Khronos Group. Information about Vulkan and the Khronos Group can be found at the Khronos Vulkan Site.

Nsight Systems can capture information about Vulkan usage by the profiled process. This includes capturing the execution time of Vulkan API functions, corresponding GPU workloads, debug util labels, and frame durations. Vulkan profiling is supported on both Windows and x86 Linux operating systems.

The Command Buffer Creation row displays time periods when command buffers were being created. This enables developers to improve their application’s multi-threaded command buffer creation. Command buffer creation time period is measured between the call to vkBeginCommandBuffer and the call to vkEndCommandBuffer.

A Queue row is displayed for each Vulkan queue created by the profiled application. The API sub-row displays time periods where vkQueueSubmit was called. The GPU Workload sub-row displays time periods where workloads were executed by the GPU.

In addition, you can see Vulkan debug util labels on both the CPU and the GPU.

Clicking on a GPU workload highlights the corresponding vkQueueSubmit call, and vice versa.

The Vulkan Memory Operations row contains an aggregation of all the Vulkan host-side memory operations, such as host-blocking writes and reads or non-persistent map-unmap ranges.

The row is separated into sub-rows by heap index and memory type - the tooltip for each row and the ranges inside show the heap flags and the memory property flags.

When tracing target application calls to Vulkan pipeline creation APIs, Nsight Systems leverages the Pipeline Creation Feedback extension to collect more details about the duration of individual pipeline creation stages.

See Pipeline Creation Feedback extension for details about this extension.

Vulkan pipeline creation feedback is available on NVIDIA driver release 435 or later.

The Frame Duration section displays frame durations on both the CPU and the GPU.

The stutter row highlights frames that are significantly longer than the other frames in their immediate vicinity.

The stutter row uses an algorithm that compares the duration of each frame to the median duration of the surrounding 19 frames. Duration difference under 4 milliseconds is never considered a stutter, to avoid cluttering the display with frames whose absolute stutter is small and not noticeable to the user.

OSC detection

The "19 frame window median" algorithm by itself may not work well with some cases of "oscillation" (consecutive fast and slow frames), resulting in some false positives. The median duration is not meaningful in cases of oscillation and can be misleading.

Right-clicking the Frame Duration row caption lets you choose the target frame rate (30, 60, 90 or custom frames per second).

By clicking the Customize FPS Display option, a customization dialog pops up. In the dialog, you can now define the frame duration threshold to customize the view of the potentially problematic frames. In addition, you can define the threshold for the stutter analysis frames.

The timing of the actual calls to the frame boundary calls can be seen in the blue bar at the bottom of the CPU frame duration row

Reflex SDK

NVIDIA Reflex SDK is a series of NVAPI calls that allow applications to integrate the Ultra Low Latency driver feature more directly into their game to further optimize synchronization between simulation and rendering stages and lower the latency between user input and final image rendering. For more details about Reflex SDK, see Reflex SDK Site.

Nsight Systems will automatically capture NVAPI functions when either Direct3D 11, Direct3D 12, or Vulkan API trace are enabled.

Performance Warnings row

The Frame Health row displays actions that took significantly a longer time during the current frame, compared to the median time of the same actions executed during the surrounding 19-frames. This is a great tool for detecting the reason for frame time stuttering. Such actions may be: shader compilation, present, memory mapping, and more. Nsight Systems measures the accumulated time of such actions in each frame. For example: calculating the accumulated time of shader compilations in each frame and comparing it to the accumulated time of shader compilations in the surrounding 19 frames.

Example of a Vulkan frame health row:

The Memory Utilization row displays the amount of used local GPU memory and the commit limit for each GPU.

Note that this is not the same as the CUDA kernel memory allocation graph, see CUDA GPU Memory Graph for that functionality.

The VSYNC rows display when the monitor's vertical synchronizations occur.

The functions listed below receive a special treatment. If the tool detects that the resource is already acquired by another thread and will induce a blocking call, we always trace it. Otherwise, it will never be traced.

pthread_mutex_lock
pthread_rwlock_rdlock
pthread_rwlock_wrlock
pthread_spin_lock
sem_wait

Note that even if a call is determined as potentially blocking, there is a chance that it may not actually block after a few cycles have elapsed. The call will still be traced in this scenario.

• Nsight Systems only traces syscall wrappers exposed by the C runtime. It is not able to trace syscall invoked through assembly code.

• Additional thread states, as well as backtrace collection on long calls, are only enabled if sampling is turned on.

• It is not possible to configure the depth and duration threshold when collecting backtraces. Currently, only OS runtime libraries calls longer than 80 μs will generate a backtrace with a maximum of 24 frames. This limitation will be removed in a future version of the product.

• It is required to compile your application and libraries with the -funwind-tables compiler flag in order for Nsight Systems to unwind the backtraces correctly.

The OS runtime libraries tracing is limited to a select list of functions. It also depends on the version of the C runtime linked to the application.

Libc system call wrappers

accept
accept4
acct
alarm
arch_prctl
bind
bpf
brk
chroot
clock_nanosleep
connect
copy_file_range
creat
creat64
dup
dup2
dup3
epoll_ctl
epoll_pwait
epoll_wait
fallocate
fallocate64
fcntl
fdatasync
flock
fork
fsync
ftruncate
futex
ioctl
ioperm
iopl
kill
killpg
listen
membarrier
mlock
mlock2
mlockall
mmap
mmap64
mount
move_pages
mprotect
mq_notify
mq_open
mq_receive
mq_send
mq_timedreceive
mq_timedsend
mremap
msgctl
msgget
msgrcv
msgsnd
msync
munmap
nanosleep
nfsservctl
open
open64
openat
openat64
pause
pipe
pipe2
pivot_root
poll
ppoll
prctl
pread
pread64
preadv
preadv2
preadv64
process_vm_readv
process_vm_writev
pselect6
ptrace
pwrite
pwrite64
pwritev
pwritev2
pwritev64
read
readv
reboot
recv
recvfrom
recvmmsg
recvmsg
rt_sigaction
rt_sigqueueinfo
rt_sigsuspend
rt_sigtimedwait
sched_yield
seccomp
select
semctl
semget
semop
semtimedop
send
sendfile
sendfile64
sendmmsg
sendmsg
sendto
shmat
shmctl
shmdt
shmget
shutdown
sigaction
sigsuspend
sigtimedwait
socket
socketpair
splice
swapoff
swapon
sync
sync_file_range
syncfs
tee
tgkill
tgsigqueueinfo
tkill
truncate
umount2
unshare
uselib
vfork
vhangup
vmsplice
wait
wait3
wait4
waitid
waitpid
write
writev
_sysctl

POSIX Threads

pthread_barrier_wait
pthread_cancel
pthread_cond_broadcast
pthread_cond_signal
pthread_cond_timedwait
pthread_cond_wait
pthread_create
pthread_join
pthread_kill
pthread_mutex_lock
pthread_mutex_timedlock
pthread_mutex_trylock
pthread_rwlock_rdlock
pthread_rwlock_timedrdlock
pthread_rwlock_timedwrlock
pthread_rwlock_tryrdlock
pthread_rwlock_trywrlock
pthread_rwlock_wrlock
pthread_spin_lock
pthread_spin_trylock
pthread_timedjoin_np
pthread_tryjoin_np
pthread_yield
sem_timedwait
sem_trywait
sem_wait

I/O

aio_fsync
aio_fsync64
aio_suspend
aio_suspend64
fclose
fcloseall
fflush
fflush_unlocked
fgetc
fgetc_unlocked
fgets
fgets_unlocked
fgetwc
fgetwc_unlocked
fgetws
fgetws_unlocked
flockfile
fopen
fopen64
fputc
fputc_unlocked
fputs
fputs_unlocked
fputwc
fputwc_unlocked
fputws
fputws_unlocked
fread
fread_unlocked
freopen
freopen64
ftrylockfile
fwrite
fwrite_unlocked
getc
getc_unlocked
getdelim
getline
getw
getwc
getwc_unlocked
lockf
lockf64
mkfifo
mkfifoat
posix_fallocate
posix_fallocate64
putc
putc_unlocked
putwc
putwc_unlocked

Miscellaneous

forkpty
popen
posix_spawn
posix_spawnp
sigwait
sigwaitinfo
sleep
system
usleep

When the Collect GPU Memory Usage option is selected from the Collect CUDA trace option set, Nsight Systems will track CUDA GPU memory allocations and deallocations and present a graph of this information in the timeline. This is not the same as the GPU memory graph generated during stutter analysis on the Windows target (see Stutter Memory Trace)

Below, in the report on the left, memory is allocated and freed during the collection. In the report on the right, memory is allocated, but not freed during the collection.

Here is another example, where allocations are happening on multiple GPUs

For Nsight Systems Workstation Edition, Unified Memory (also called Managed Memory) transfer trace is enabled automatically in Nsight Systems when CUDA trace is selected. It incurs no overhead in programs that do not perform any Unified Memory transfers. Data is displayed in the Managed Memory area of the timeline:

HtoD transfer indicates the CUDA kernel accessed managed memory that was residing on the host, so the kernel execution paused and transferred the data to the device. Heavy traffic here will incur performance penalties in CUDA kernels, so consider using manual cudaMemcpy operations from pinned host memory instead.

PtoP transfer indicates the CUDA kernel accessed managed memory that was residing on a different device, so the kernel execution paused and transferred the data to this device. Heavy traffic here will incur performance penalties, so consider using manual cudaMemcpyPeer operations to transfer from other devices' memory instead. The row showing these events is for the destination device -- the source device is shown in the tooltip for each transfer event.

DtoH transfer indicates the CPU accessed managed memory that was residing on a CUDA device, so the CPU execution paused and transferred the data to system memory. Heavy traffic here will incur performance penalties in CPU code, so consider using manual cudaMemcpy operations from pinned host memory instead.

Some Unified Memory transfers are highlighted with red to indicate potential performance issues:

Nsight Systems is capable of capturing information about CUDA graphs in your application at either the graph or node granularity. This can be set in the CLI using the --cuda-graph-trace option, or in the GUI by setting the appropriate drop down.

When CUDA graph trace is set to graph, the users sees each graph as one item on the timeline:

When CUDA graph trace is set to node, the users sees each graph as a set of nodes on the timeline:

Tracing CUDA graphs at the graph level rather than the tracing the underlying nodes results in significantly less overhead. This option is only available with CUDA driver 515.43 or higher.

CUDA Runtime API

cudaBindSurfaceToArray
cudaBindTexture
cudaBindTexture2D
cudaBindTextureToArray
cudaBindTextureToMipmappedArray
cudaConfigureCall
cudaCreateSurfaceObject
cudaCreateTextureObject
cudaD3D10MapResources
cudaD3D10RegisterResource
cudaD3D10UnmapResources
cudaD3D10UnregisterResource
cudaD3D9MapResources
cudaD3D9MapVertexBuffer
cudaD3D9RegisterResource
cudaD3D9RegisterVertexBuffer
cudaD3D9UnmapResources
cudaD3D9UnmapVertexBuffer
cudaD3D9UnregisterResource
cudaD3D9UnregisterVertexBuffer
cudaDestroySurfaceObject
cudaDestroyTextureObject
cudaDeviceReset
cudaDeviceSynchronize
cudaEGLStreamConsumerAcquireFrame
cudaEGLStreamConsumerConnect
cudaEGLStreamConsumerConnectWithFlags
cudaEGLStreamConsumerDisconnect
cudaEGLStreamConsumerReleaseFrame
cudaEGLStreamConsumerReleaseFrame
cudaEGLStreamProducerConnect
cudaEGLStreamProducerDisconnect
cudaEGLStreamProducerReturnFrame
cudaEventCreate
cudaEventCreateFromEGLSync
cudaEventCreateWithFlags
cudaEventDestroy
cudaEventQuery
cudaEventRecord
cudaEventRecord_ptsz
cudaEventSynchronize
cudaFree
cudaFreeArray
cudaFreeHost
cudaFreeMipmappedArray
cudaGLMapBufferObject
cudaGLMapBufferObjectAsync
cudaGLRegisterBufferObject
cudaGLUnmapBufferObject
cudaGLUnmapBufferObjectAsync
cudaGLUnregisterBufferObject
cudaGraphicsD3D10RegisterResource
cudaGraphicsD3D11RegisterResource
cudaGraphicsD3D9RegisterResource
cudaGraphicsEGLRegisterImage
cudaGraphicsGLRegisterBuffer
cudaGraphicsGLRegisterImage
cudaGraphicsMapResources
cudaGraphicsUnmapResources
cudaGraphicsUnregisterResource
cudaGraphicsVDPAURegisterOutputSurface
cudaGraphicsVDPAURegisterVideoSurface
cudaHostAlloc
cudaHostRegister
cudaHostUnregister
cudaLaunch
cudaLaunchCooperativeKernel
cudaLaunchCooperativeKernelMultiDevice
cudaLaunchCooperativeKernel_ptsz
cudaLaunchKernel
cudaLaunchKernel_ptsz
cudaLaunch_ptsz
cudaMalloc
cudaMalloc3D
cudaMalloc3DArray
cudaMallocArray
cudaMallocHost
cudaMallocManaged
cudaMallocMipmappedArray
cudaMallocPitch
cudaMemGetInfo
cudaMemPrefetchAsync
cudaMemPrefetchAsync_ptsz
cudaMemcpy
cudaMemcpy2D
cudaMemcpy2DArrayToArray
cudaMemcpy2DArrayToArray_ptds
cudaMemcpy2DAsync
cudaMemcpy2DAsync_ptsz
cudaMemcpy2DFromArray
cudaMemcpy2DFromArrayAsync
cudaMemcpy2DFromArrayAsync_ptsz
cudaMemcpy2DFromArray_ptds
cudaMemcpy2DToArray
cudaMemcpy2DToArrayAsync
cudaMemcpy2DToArrayAsync_ptsz
cudaMemcpy2DToArray_ptds
cudaMemcpy2D_ptds
cudaMemcpy3D
cudaMemcpy3DAsync
cudaMemcpy3DAsync_ptsz
cudaMemcpy3DPeer
cudaMemcpy3DPeerAsync
cudaMemcpy3DPeerAsync_ptsz
cudaMemcpy3DPeer_ptds
cudaMemcpy3D_ptds
cudaMemcpyArrayToArray
cudaMemcpyArrayToArray_ptds
cudaMemcpyAsync
cudaMemcpyAsync_ptsz
cudaMemcpyFromArray
cudaMemcpyFromArrayAsync
cudaMemcpyFromArrayAsync_ptsz
cudaMemcpyFromArray_ptds
cudaMemcpyFromSymbol
cudaMemcpyFromSymbolAsync
cudaMemcpyFromSymbolAsync_ptsz
cudaMemcpyFromSymbol_ptds
cudaMemcpyPeer
cudaMemcpyPeerAsync
cudaMemcpyToArray
cudaMemcpyToArrayAsync
cudaMemcpyToArrayAsync_ptsz
cudaMemcpyToArray_ptds
cudaMemcpyToSymbol
cudaMemcpyToSymbolAsync
cudaMemcpyToSymbolAsync_ptsz
cudaMemcpyToSymbol_ptds
cudaMemcpy_ptds
cudaMemset
cudaMemset2D
cudaMemset2DAsync
cudaMemset2DAsync_ptsz
cudaMemset2D_ptds
cudaMemset3D
cudaMemset3DAsync
cudaMemset3DAsync_ptsz
cudaMemset3D_ptds
cudaMemsetAsync
cudaMemsetAsync_ptsz
cudaMemset_ptds
cudaPeerRegister
cudaPeerUnregister
cudaStreamAddCallback
cudaStreamAddCallback_ptsz
cudaStreamAttachMemAsync
cudaStreamAttachMemAsync_ptsz
cudaStreamCreate
cudaStreamCreateWithFlags
cudaStreamCreateWithPriority
cudaStreamDestroy
cudaStreamQuery
cudaStreamQuery_ptsz
cudaStreamSynchronize
cudaStreamSynchronize_ptsz
cudaStreamWaitEvent
cudaStreamWaitEvent_ptsz
cudaThreadSynchronize
cudaUnbindTexture

CUDA Primary API

cu64Array3DCreate
cu64ArrayCreate
cu64D3D9MapVertexBuffer
cu64GLMapBufferObject
cu64GLMapBufferObjectAsync
cu64MemAlloc
cu64MemAllocPitch
cu64MemFree
cu64MemGetInfo
cu64MemHostAlloc
cu64Memcpy2D
cu64Memcpy2DAsync
cu64Memcpy2DUnaligned
cu64Memcpy3D
cu64Memcpy3DAsync
cu64MemcpyAtoD
cu64MemcpyDtoA
cu64MemcpyDtoD
cu64MemcpyDtoDAsync
cu64MemcpyDtoH
cu64MemcpyDtoHAsync
cu64MemcpyHtoD
cu64MemcpyHtoDAsync
cu64MemsetD16
cu64MemsetD16Async
cu64MemsetD2D16
cu64MemsetD2D16Async
cu64MemsetD2D32
cu64MemsetD2D32Async
cu64MemsetD2D8
cu64MemsetD2D8Async
cu64MemsetD32
cu64MemsetD32Async
cu64MemsetD8
cu64MemsetD8Async
cuArray3DCreate
cuArray3DCreate_v2
cuArrayCreate
cuArrayCreate_v2
cuArrayDestroy
cuBinaryFree
cuCompilePtx
cuCtxCreate
cuCtxCreate_v2
cuCtxDestroy
cuCtxDestroy_v2
cuCtxSynchronize
cuD3D10CtxCreate
cuD3D10CtxCreateOnDevice
cuD3D10CtxCreate_v2
cuD3D10MapResources
cuD3D10RegisterResource
cuD3D10UnmapResources
cuD3D10UnregisterResource
cuD3D11CtxCreate
cuD3D11CtxCreateOnDevice
cuD3D11CtxCreate_v2
cuD3D9CtxCreate
cuD3D9CtxCreateOnDevice
cuD3D9CtxCreate_v2
cuD3D9MapResources
cuD3D9MapVertexBuffer
cuD3D9MapVertexBuffer_v2
cuD3D9RegisterResource
cuD3D9RegisterVertexBuffer
cuD3D9UnmapResources
cuD3D9UnmapVertexBuffer
cuD3D9UnregisterResource
cuD3D9UnregisterVertexBuffer
cuEGLStreamConsumerAcquireFrame
cuEGLStreamConsumerConnect
cuEGLStreamConsumerConnectWithFlags
cuEGLStreamConsumerDisconnect
cuEGLStreamConsumerReleaseFrame
cuEGLStreamProducerConnect
cuEGLStreamProducerDisconnect
cuEGLStreamProducerPresentFrame
cuEGLStreamProducerReturnFrame
cuEventCreate
cuEventCreateFromEGLSync
cuEventCreateFromNVNSync
cuEventDestroy
cuEventDestroy_v2
cuEventQuery
cuEventRecord
cuEventRecord_ptsz
cuEventSynchronize
cuGLCtxCreate
cuGLCtxCreate_v2
cuGLInit
cuGLMapBufferObject
cuGLMapBufferObjectAsync
cuGLMapBufferObjectAsync_v2
cuGLMapBufferObjectAsync_v2_ptsz
cuGLMapBufferObject_v2
cuGLMapBufferObject_v2_ptds
cuGLRegisterBufferObject
cuGLUnmapBufferObject
cuGLUnmapBufferObjectAsync
cuGLUnregisterBufferObject
cuGraphicsD3D10RegisterResource
cuGraphicsD3D11RegisterResource
cuGraphicsD3D9RegisterResource
cuGraphicsEGLRegisterImage
cuGraphicsGLRegisterBuffer
cuGraphicsGLRegisterImage
cuGraphicsMapResources
cuGraphicsMapResources_ptsz
cuGraphicsUnmapResources
cuGraphicsUnmapResources_ptsz
cuGraphicsUnregisterResource
cuGraphicsVDPAURegisterOutputSurface
cuGraphicsVDPAURegisterVideoSurface
cuInit
cuLaunch
cuLaunchCooperativeKernel
cuLaunchCooperativeKernelMultiDevice
cuLaunchCooperativeKernel_ptsz
cuLaunchGrid
cuLaunchGridAsync
cuLaunchKernel
cuLaunchKernel_ptsz
cuLinkComplete
cuLinkCreate
cuLinkCreate_v2
cuLinkDestroy
cuMemAlloc
cuMemAllocHost
cuMemAllocHost_v2
cuMemAllocManaged
cuMemAllocPitch
cuMemAllocPitch_v2
cuMemAlloc_v2
cuMemFree
cuMemFreeHost
cuMemFree_v2
cuMemGetInfo
cuMemGetInfo_v2
cuMemHostAlloc
cuMemHostAlloc_v2
cuMemHostRegister
cuMemHostRegister_v2
cuMemHostUnregister
cuMemPeerRegister
cuMemPeerUnregister
cuMemPrefetchAsync
cuMemPrefetchAsync_ptsz
cuMemcpy
cuMemcpy2D
cuMemcpy2DAsync
cuMemcpy2DAsync_v2
cuMemcpy2DAsync_v2_ptsz
cuMemcpy2DUnaligned
cuMemcpy2DUnaligned_v2
cuMemcpy2DUnaligned_v2_ptds
cuMemcpy2D_v2
cuMemcpy2D_v2_ptds
cuMemcpy3D
cuMemcpy3DAsync
cuMemcpy3DAsync_v2
cuMemcpy3DAsync_v2_ptsz
cuMemcpy3DPeer
cuMemcpy3DPeerAsync
cuMemcpy3DPeerAsync_ptsz
cuMemcpy3DPeer_ptds
cuMemcpy3D_v2
cuMemcpy3D_v2_ptds
cuMemcpyAsync
cuMemcpyAsync_ptsz
cuMemcpyAtoA
cuMemcpyAtoA_v2
cuMemcpyAtoA_v2_ptds
cuMemcpyAtoD
cuMemcpyAtoD_v2
cuMemcpyAtoD_v2_ptds
cuMemcpyAtoH
cuMemcpyAtoHAsync
cuMemcpyAtoHAsync_v2
cuMemcpyAtoHAsync_v2_ptsz
cuMemcpyAtoH_v2
cuMemcpyAtoH_v2_ptds
cuMemcpyDtoA
cuMemcpyDtoA_v2
cuMemcpyDtoA_v2_ptds
cuMemcpyDtoD
cuMemcpyDtoDAsync
cuMemcpyDtoDAsync_v2
cuMemcpyDtoDAsync_v2_ptsz
cuMemcpyDtoD_v2
cuMemcpyDtoD_v2_ptds
cuMemcpyDtoH
cuMemcpyDtoHAsync
cuMemcpyDtoHAsync_v2
cuMemcpyDtoHAsync_v2_ptsz
cuMemcpyDtoH_v2
cuMemcpyDtoH_v2_ptds
cuMemcpyHtoA
cuMemcpyHtoAAsync
cuMemcpyHtoAAsync_v2
cuMemcpyHtoAAsync_v2_ptsz
cuMemcpyHtoA_v2
cuMemcpyHtoA_v2_ptds
cuMemcpyHtoD
cuMemcpyHtoDAsync
cuMemcpyHtoDAsync_v2
cuMemcpyHtoDAsync_v2_ptsz
cuMemcpyHtoD_v2
cuMemcpyHtoD_v2_ptds
cuMemcpyPeer
cuMemcpyPeerAsync
cuMemcpyPeerAsync_ptsz
cuMemcpyPeer_ptds
cuMemcpy_ptds
cuMemcpy_v2
cuMemsetD16
cuMemsetD16Async
cuMemsetD16Async_ptsz
cuMemsetD16_v2
cuMemsetD16_v2_ptds
cuMemsetD2D16
cuMemsetD2D16Async
cuMemsetD2D16Async_ptsz
cuMemsetD2D16_v2
cuMemsetD2D16_v2_ptds
cuMemsetD2D32
cuMemsetD2D32Async
cuMemsetD2D32Async_ptsz
cuMemsetD2D32_v2
cuMemsetD2D32_v2_ptds
cuMemsetD2D8
cuMemsetD2D8Async
cuMemsetD2D8Async_ptsz
cuMemsetD2D8_v2
cuMemsetD2D8_v2_ptds
cuMemsetD32
cuMemsetD32Async
cuMemsetD32Async_ptsz
cuMemsetD32_v2
cuMemsetD32_v2_ptds
cuMemsetD8
cuMemsetD8Async
cuMemsetD8Async_ptsz
cuMemsetD8_v2
cuMemsetD8_v2_ptds
cuMipmappedArrayCreate
cuMipmappedArrayDestroy
cuModuleLoad
cuModuleLoadData
cuModuleLoadDataEx
cuModuleLoadFatBinary
cuModuleUnload
cuStreamAddCallback
cuStreamAddCallback_ptsz
cuStreamAttachMemAsync
cuStreamAttachMemAsync_ptsz
cuStreamBatchMemOp
cuStreamBatchMemOp_ptsz
cuStreamCreate
cuStreamCreateWithPriority
cuStreamDestroy
cuStreamDestroy_v2
cuStreamSynchronize
cuStreamSynchronize_ptsz
cuStreamWaitEvent
cuStreamWaitEvent_ptsz
cuStreamWaitValue32
cuStreamWaitValue32_ptsz
cuStreamWaitValue64
cuStreamWaitValue64_ptsz
cuStreamWriteValue32
cuStreamWriteValue32_ptsz
cuStreamWriteValue64
cuStreamWriteValue64_ptsz
cuSurfObjectCreate
cuSurfObjectDestroy
cuSurfRefCreate
cuSurfRefDestroy
cuTexObjectCreate
cuTexObjectDestroy
cuTexRefCreate
cuTexRefDestroy
cuVDPAUCtxCreate
cuVDPAUCtxCreate_v2

cuDNN API functions

cudnnActivationBackward
cudnnActivationBackward_v3
cudnnActivationBackward_v4
cudnnActivationForward
cudnnActivationForward_v3
cudnnActivationForward_v4
cudnnAddTensor
cudnnBatchNormalizationBackward
cudnnBatchNormalizationBackwardEx
cudnnBatchNormalizationForwardInference
cudnnBatchNormalizationForwardTraining
cudnnBatchNormalizationForwardTrainingEx
cudnnCTCLoss
cudnnConvolutionBackwardBias
cudnnConvolutionBackwardData
cudnnConvolutionBackwardFilter
cudnnConvolutionBiasActivationForward
cudnnConvolutionForward
cudnnCreate
cudnnCreateAlgorithmPerformance
cudnnDestroy
cudnnDestroyAlgorithmPerformance
cudnnDestroyPersistentRNNPlan
cudnnDivisiveNormalizationBackward
cudnnDivisiveNormalizationForward
cudnnDropoutBackward
cudnnDropoutForward
cudnnDropoutGetReserveSpaceSize
cudnnDropoutGetStatesSize
cudnnFindConvolutionBackwardDataAlgorithm
cudnnFindConvolutionBackwardDataAlgorithmEx
cudnnFindConvolutionBackwardFilterAlgorithm
cudnnFindConvolutionBackwardFilterAlgorithmEx
cudnnFindConvolutionForwardAlgorithm
cudnnFindConvolutionForwardAlgorithmEx
cudnnFindRNNBackwardDataAlgorithmEx
cudnnFindRNNBackwardWeightsAlgorithmEx
cudnnFindRNNForwardInferenceAlgorithmEx
cudnnFindRNNForwardTrainingAlgorithmEx
cudnnFusedOpsExecute
cudnnIm2Col
cudnnLRNCrossChannelBackward
cudnnLRNCrossChannelForward
cudnnMakeFusedOpsPlan
cudnnMultiHeadAttnBackwardData
cudnnMultiHeadAttnBackwardWeights
cudnnMultiHeadAttnForward
cudnnOpTensor
cudnnPoolingBackward
cudnnPoolingForward
cudnnRNNBackwardData
cudnnRNNBackwardDataEx
cudnnRNNBackwardWeights
cudnnRNNBackwardWeightsEx
cudnnRNNForwardInference
cudnnRNNForwardInferenceEx
cudnnRNNForwardTraining
cudnnRNNForwardTrainingEx
cudnnReduceTensor
cudnnReorderFilterAndBias
cudnnRestoreAlgorithm
cudnnRestoreDropoutDescriptor
cudnnSaveAlgorithm
cudnnScaleTensor
cudnnSoftmaxBackward
cudnnSoftmaxForward
cudnnSpatialTfGridGeneratorBackward
cudnnSpatialTfGridGeneratorForward
cudnnSpatialTfSamplerBackward
cudnnSpatialTfSamplerForward
cudnnTransformFilter
cudnnTransformTensor
cudnnTransformTensorEx

For general information on using the target CLI, see CLI Profiling on Linux. For the CLI, the functions that are traced are set to the following list:

glWaitSync
glReadPixels
glReadnPixelsKHR
glReadnPixelsEXT
glReadnPixelsARB
glReadnPixels
glFlush
glFinishFenceNV
glFinish
glClientWaitSync
glClearTexSubImage
glClearTexImage
glClearStencil
glClearNamedFramebufferuiv
glClearNamedFramebufferiv
glClearNamedFramebufferfv
glClearNamedFramebufferfi
glClearNamedBufferSubDataEXT
glClearNamedBufferSubData
glClearNamedBufferDataEXT
glClearNamedBufferData
glClearIndex
glClearDepthx
glClearDepthf
glClearDepthdNV
glClearDepth
glClearColorx
glClearColorIuiEXT
glClearColorIiEXT
glClearColor
glClearBufferuiv
glClearBufferSubData
glClearBufferiv
glClearBufferfv
glClearBufferfi
glClearBufferData
glClearAccum
glClear
glDispatchComputeIndirect
glDispatchComputeGroupSizeARB
glDispatchCompute
glComputeStreamNV
glNamedFramebufferDrawBuffers
glNamedFramebufferDrawBuffer
glMultiDrawElementsIndirectEXT
glMultiDrawElementsIndirectCountARB
glMultiDrawElementsIndirectBindlessNV
glMultiDrawElementsIndirectBindlessCountNV
glMultiDrawElementsIndirectAMD
glMultiDrawElementsIndirect
glMultiDrawElementsEXT
glMultiDrawElementsBaseVertex
glMultiDrawElements
glMultiDrawArraysIndirectEXT
glMultiDrawArraysIndirectCountARB
glMultiDrawArraysIndirectBindlessNV
glMultiDrawArraysIndirectBindlessCountNV
glMultiDrawArraysIndirectAMD
glMultiDrawArraysIndirect
glMultiDrawArraysEXT
glMultiDrawArrays
glListDrawCommandsStatesClientNV
glFramebufferDrawBuffersEXT
glFramebufferDrawBufferEXT
glDrawTransformFeedbackStreamInstanced
glDrawTransformFeedbackStream
glDrawTransformFeedbackNV
glDrawTransformFeedbackInstancedEXT
glDrawTransformFeedbackInstanced
glDrawTransformFeedbackEXT
glDrawTransformFeedback
glDrawTexxvOES
glDrawTexxOES
glDrawTextureNV
glDrawTexsvOES
glDrawTexsOES
glDrawTexivOES
glDrawTexiOES
glDrawTexfvOES
glDrawTexfOES
glDrawRangeElementsEXT
glDrawRangeElementsBaseVertexOES
glDrawRangeElementsBaseVertexEXT
glDrawRangeElementsBaseVertex
glDrawRangeElements
glDrawPixels
glDrawElementsInstancedNV
glDrawElementsInstancedEXT
glDrawElementsInstancedBaseVertexOES
glDrawElementsInstancedBaseVertexEXT
glDrawElementsInstancedBaseVertexBaseInstanceEXT
glDrawElementsInstancedBaseVertexBaseInstance
glDrawElementsInstancedBaseVertex
glDrawElementsInstancedBaseInstanceEXT
glDrawElementsInstancedBaseInstance
glDrawElementsInstancedARB
glDrawElementsInstanced
glDrawElementsIndirect
glDrawElementsBaseVertexOES
glDrawElementsBaseVertexEXT
glDrawElementsBaseVertex
glDrawElements
glDrawCommandsStatesNV
glDrawCommandsStatesAddressNV
glDrawCommandsNV
glDrawCommandsAddressNV
glDrawBuffersNV
glDrawBuffersATI
glDrawBuffersARB
glDrawBuffers
glDrawBuffer
glDrawArraysInstancedNV
glDrawArraysInstancedEXT
glDrawArraysInstancedBaseInstanceEXT
glDrawArraysInstancedBaseInstance
glDrawArraysInstancedARB
glDrawArraysInstanced
glDrawArraysIndirect
glDrawArraysEXT
glDrawArrays
eglSwapBuffersWithDamageKHR
eglSwapBuffers
glXSwapBuffers
glXQueryDrawable
glXGetCurrentReadDrawable
glXGetCurrentDrawable
glGetQueryObjectuivEXT
glGetQueryObjectuivARB
glGetQueryObjectuiv
glGetQueryObjectivARB
glGetQueryObjectiv

GPU Metrics feature is intended to identify performance limiters in applications using GPU for computations and graphics. It uses periodic sampling to gather performance metrics and detailed timing statistics associated with different GPU hardware units taking advantage of specialized hardware to capture this data in a single pass with minimal overhead.

Note: GPU Metrics will give you precise device level information, but it does not know which process or context is involved. GPU context switch trace provides less precise information, but will give you process and context information.

These metrics provide an overview of GPU efficiency over time within compute, graphics, and input/output (IO) activities such as:

• IO throughputs: PCIe, NVLink, and GPU memory bandwidth
• SM utilization: SMs activity, tensor core activity, instructions issued, warp occupancy, and unassigned warp slots

It is designed to help users answer the common questions:

• Is my GPU idle?
• Is my GPU full? Enough kernel grids size and streams? Are my SMs and warp slots full?
• Am I using TensorCores?
• Is my instruction rate high?
• Am I possibly blocked on IO, or number of warps, etc

Nsight Systems GPU Metrics is only available for Linux targets on x86-64 and aarch64, and for Windows targets. It requires NVIDIA Turing architecture or newer.

Note: Elevated permissions are required. On Linux use sudo to elevate privileges. On Windows the user must run from an admin command prompt or accept the UAC escalation dialog. See Permissions Issues and Performance Counters for more information.

For commands to launch GPU Metrics from the CLI with examples, see the CLI documentation.

When launching analysis in Nsight Systems, select Collect GPU Metrics.

Select the GPUs dropdown to pick which GPUs you wish to sample.

Select the Metric set: dropdown to choose which available metric set you would like to sample.

Note that metric sets for GPUs that are not being sampled will be greyed out.

Sampling frequency can be selected from the range of 10 Hz - 200 kHz. The default value is 10 kHz.

The maximum sampling frequency without buffer overflow events depends on GPU (SM count), GPU load intensity, and overall system load. The bigger the chip and the higher the load, the lower the maximum frequency. If you need higher frequency, you can increase it until you get "Buffer overflow" message in the Diagnostics Summary report page.

Each metric set has a recommended sampling frequency range in its description. These ranges are approximate. If you observe Inonsistent Data or Missing Data ranges on timeline, please try closer to the recommended frequency.

• GPC Clock Frequency - gpc__cycles_elapsed.avg.per_second

  The average GPC clock frequency in hertz. In public documentation the GPC clock may be called the "Application" clock, "Graphic" clock, "Base" clock, or "Boost" clock.

  Note: The collection mechanism for GPC can result in a small fluctuation between samples.

• SYS Clock Frequency - sys__cycles_elapsed.avg.per_second

  The average SYS clock frequency in hertz. The GPU front end (command processor), copy engines, and the performance monitor run at the SYS clock. On Turing and NVIDIA GA100 GPUs the sampling frequency is based upon a period of SYS clocks (not time) so samples per second will vary with SYS clock. On NVIDIA GA10x GPUs the sampling frequency is based upon a fixed frequency clock. The maximum frequency scales linearly with the SYS clock.

• GR Active - gr__cycles_active.sum.pct_of_peak_sustained_elapsed

  The percentage of cycles the graphics/compute engine is active. The graphics/compute engine is active if there is any work in the graphics pipe or if the compute pipe is processing work.

  GA100 MIG - MIG is not yet supported. This counter will report the activity of the primary GR engine.

• Sync Compute In Flight - gr__dispatch_cycles_active_queue_sync.avg.pct_of_peak_sustained_elapsed

  The percentage of cycles with synchronous compute in flight.

  CUDA: CUDA will only report synchronous queue in the case of MPS configured with 64 sub-context. Synchronous refers to work submitted in VEID=0.

  Graphics: This will be true if any compute work submitted from the direct queue is in flight.

• Async Compute in Flight - gr__dispatch_cycles_active_queue_async.avg.pct_of_peak_sustained_elapsed

  The percentage of cycles with asynchronous compute in flight.

  CUDA: CUDA will only report all compute work as asynchronous. The one exception is if MPS is configured and all 64 sub-context are in use. 1 sub-context (VEID=0) will report as synchronous.

  Graphics: This will be true if any compute work submitted from a compute queue is in flight.

• Draw Started - fe__draw_count.avg.pct_of_peak_sustained_elapsed

  The ratio of draw calls issued to the graphics pipe to the maximum sustained rate of the graphics pipe.

  Note:The percentage will always be very low as the front end can issue draw calls significantly faster than the pipe can execute the draw call. The rendering of this row will be changed to help indicate when draw calls are being issued.

• Dispatch Started - gr__dispatch_count.avg.pct_of_peak_sustained_elapsed

  The ratio of compute grid launches (dispatches) to the compute pipe to the maximum sustained rate of the compute pipe.

  Note: The percentage will always be very low as the front end can issue grid launches significantly faster than the pipe can execute the draw call. The rendering of this row will be changed to help indicate when grid launches are being issued.

• Vertex/Tess/Geometry Warps in Flight - tpc__warps_active_shader_vtg_realtime.avg.pct_of_peak_sustained_elapsed

  The ratio of active vertex, geometry, tessellation, and meshlet shader warps resident on the SMs to the maximum number of warps per SM as a percentage.

• Pixel Warps in Flight - tpc__warps_active_shader_ps_realtime.avg.pct_of_peak_sustained_elapsed

  The ratio of active pixel/fragment shader warps resident on the SMs to the maximum number of warps per SM as a percentage.

• Compute Warps in Flight - tpc__warps_active_shader_cs_realtime.avg.pct_of_peak_sustained_elapsed

  The ratio of active compute shader warps resident on the SMs to the maximum number of warps per SM as a percentage.

• Active SM Unused Warp Slots - tpc__warps_inactive_sm_active_realtime.avg.pct_of_peak_sustained_elapsed

  The ratio of inactive warp slots on the SMs to the maximum number of warps per SM as a percentage. This is an indication of how many more warps may fit on the SMs if occupancy is not limited by a resource such as max warps of a shader type, shared memory, registers per thread, or thread blocks per SM.

• Idle SM Unused Warp Slots - tpc__warps_inactive_sm_idle_realtime.avg.pct_of_peak_sustained_elapsed

  The ratio of inactive warps slots due to idle SMs to the the maximum number of warps per SM as a percentage.

  This is an indicator that the current workload on the SM is not sufficient to put work on all SMs. This can be due to:

  • CPU starving the GPU
  • current work is too small to saturate the GPU
  • current work is trailing off but blocking next work
• SM Active - sm__cycles_active.avg.pct_of_peak_sustained_elapsed

  The ratio of cycles SMs had at least 1 warp in flight (allocated on SM) to the number of cycles as a percentage. A value of 0 indicates all SMs were idle (no warps in flight). A value of 50% can indicate some gradient between all SMs active 50% of the sample period or 50% of SMs active 100% of the sample period.

• SM Issue - sm__inst_executed_realtime.avg.pct_of_peak_sustained_elapsed

  The ratio of cycles that SM sub-partitions (warp schedulers) issued an instruction to the number of cycles in the sample period as a percentage.

• Tensor Active - sm__pipe_tensor_cycles_active_realtime.avg.pct_of_peak_sustained_elapsed

  The ratio of cycles the SM tensor pipes were active issuing tensor instructions to the number of cycles in the sample period as a percentage.

  TU102/4/6: This metric is not available on TU10x for periodic sampling. Please see Tensor Active/FP16 Active.

• Tensor Active / FP16 Active - sm__pipe_shared_cycles_active_realtime.avg.pct_of_peak_sustained_elapsed

  TU102/4/6 only

  The ratio of cycles the SM tensor pipes or FP16x2 pipes were active issuing tensor instructions to the number of cycles in the sample period as a percentage.

• VRAM Bandwidth - dram__throughput.avg.pct_of_peak_sustained_elapsed

  The ratio of cycles the GPU device memory controllers were actively performing read or write operations to the number of cycles in the sample period as a percentage.

• NVLink bytes received - nvlrx__bytes.avg.pct_of_peak_sustained_elapsed

  The ratio of bytes received on the NVLink interface to the maximum number of bytes receivable in the sample period as a percentage. This value includes protocol overhead.

• NVLink bytes transmitted - nvltx__bytes.avg.pct_of_peak_sustained_elapsed

  The ratio of bytes transmitted on the NVLink interface to the maximum number of bytes transmittable in the sample period as a percentage. This value includes protocol overhead.

• PCIe Read Throughput - pcie__read_bytes.avg.pct_of_peak_sustained_elapsed

  The ratio of bytes received on the PCIe interface to the maximum number of bytes receivable in the sample period as a percentage. The theoretical value is calculated based upon the PCIe generation and number of lanes. This value includes protocol overhead.

• PCIe Write Throughput - pcie__write_bytes.avg.pct_of_peak_sustained_elapsed

  The ratio of bytes received on the PCIe interface to the maximum number of bytes receivable in the sample period as a percentage. The theoretical value is calculated based upon the PCIe generation and number of lanes. This value includes protocol overhead.

It is possible to access metric values for automated processing using the Nsight Systems CLI export capabilities.

An example that extracts values of "SM Active":

      $ nsys export -t sqlite report.nsys-rep
      $ sqlite3 report.sqlite "SELECT rawTimestamp, CAST(JSON_EXTRACT(data, '$.\"SM Active\"') as INTEGER) as value FROM GENERIC_EVENTS WHERE value != 0 LIMIT 10"

      309277039|80
      309301295|99
      309325583|99
      309349776|99
      309373872|60
      309397872|19
      309421840|100
      309446000|100
      309470096|100
      309494161|99
      

An overview of data stored in each event (JSON):

      $ sqlite3 report.sqlite "SELECT data FROM GENERIC_EVENTS LIMIT 1"
      {
      "Unallocated Warps in Active SM": "0",
      "Compute Warps In Flight": "52",
      "Pixel Warps In Flight": "0",
      "Vertex\/Tess\/Geometry Warps In Flight": "0",
      "Total SM Occupancy": "52",
      "GR Active (GE\/CE)": "100",
      "Sync Compute In Flight": "0",
      "Async Compute In Flight": "98",
      "NVLink bytes received": "0",
      "NVLink bytes transmitted": "0",
      "PCIe Rx Throughput": "0",
      "PCIe Tx Throughput": "1",
      "DRAM Read Throughput": "0",
      "DRAM Write Throughput": "0",
      "Tensor Active \/ FP16 Active": "0",
      "SM Issue": "10",
      "SM Active": "52"
      }
      

Values are integer percentages (0..100)

NvEncodeAPICreateInstance
nvEncOpenEncodeSession
nvEncGetEncodeGUIDCount
nvEncGetEncodeGUIDs
nvEncGetEncodeProfileGUIDCount
nvEncGetEncodeProfileGUIDs
nvEncGetInputFormatCount
nvEncGetInputFormats
nvEncGetEncodeCaps
nvEncGetEncodePresetCount
nvEncGetEncodePresetGUIDs
nvEncGetEncodePresetConfig
nvEncGetEncodePresetConfigEx
nvEncInitializeEncoder
nvEncCreateInputBuffer
nvEncDestroyInputBuffer
nvEncCreateBitstreamBuffer
nvEncDestroyBitstreamBuffer
nvEncEncodePicture
nvEncLockBitstream
nvEncUnlockBitstream
nvEncLockInputBuffer
nvEncUnlockInputBuffer
nvEncGetEncodeStats
nvEndGetSequenceParams
nvEncRegisterAsyncEvent
nvEncUnregisterAsyncEvent
nvEncMapInputResource
nvEncUnmapInputResource
nvEncDestroyEncoder
nvEncInvalidateRefFrames
nvEncOpenEncodeSessionEx
nvEncRegisterResource
nvEncUnregisterResource
nvEncReconfigureEncoder
nvEncCreateMVBuffer
nvEncDestroyMVBuffer
nvEncRunMotionEstimationOnly
nvEncGetLastErrorString
nvEncSetIOCudaStreams
nvEncGetSequenceParamEx
       

cuvidCreateVideoSource
cuvidCreateVideoSourceW
cuvidDestroyVideoSource
cuvidSetVideoSourceState
cudaVideoState
cuvidGetSourceVideoFormat
cuvidGetSourceAudioFormat
cuvidCreateVideoParser
cuvidParseVideoData
cuvidDestroyVideoParser
cuvidCreateDecoder
cuvidDestroyDecoder
cuvidDecodePicture
cuvidGetDecodeStatus
cuvidReconfigureDecoder
cuvidMapVideoFrame
cuvidUnmapVideoFrame
cuvidMapVideoFrame64
cuvidUnmapVideoFrame64
cuvidCtxLockCreate
cuvidCtxLockDestroy
cuvidCtxLock
cuvidCtxUnlock
       

nvjpegBufferDeviceCreate
nvjpegBufferDeviceDestroy
nvjpegBufferDeviceRetrieve
nvjpegBufferPinnedCreate
nvjpegBufferPinnedDestroy
nvjpegBufferPinnedRetrieve
nvjpegCreate
nvjpegCreateEx
nvjpegCreateSimple
nvjpegDecode
nvjpegDecodeBatched
nvjpegDecodeBatchedEx
nvjpegDecodeBatchedInitialize
nvjpegDecodeBatchedPreAllocate
nvjpegDecodeBatchedSupported
nvjpegDecodeBatchedSupportedEx
nvjpegDecodeJpeg
nvjpegDecodeJpegDevice
nvjpegDecodeJpegHost
nvjpegDecodeJpegTransferToDevice
nvjpegDecodeParamsCreate
nvjpegDecodeParamsDestroy
nvjpegDecodeParamsSetAllowCMYK
nvjpegDecodeParamsSetOutputFormat
nvjpegDecodeParamsSetROI
nvjpegDecodeParamsSetScaleFactor
nvjpegDecoderCreate
nvjpegDecoderDestroy
nvjpegDecoderJpegSupported
nvjpegDecoderStateCreate
nvjpegDestroy
nvjpegEncodeGetBufferSize
nvjpegEncodeImage
nvjpegEncodeRetrieveBitstream
nvjpegEncodeRetrieveBitstreamDevice
nvjpegEncoderParamsCopyHuffmanTables
nvjpegEncoderParamsCopyMetadata
nvjpegEncoderParamsCopyQuantizationTables
nvjpegEncoderParamsCreate
nvjpegEncoderParamsDestroy
nvjpegEncoderParamsSetEncoding
nvjpegEncoderParamsSetOptimizedHuffman
nvjpegEncoderParamsSetQuality
nvjpegEncoderParamsSetSamplingFactors
nvjpegEncoderStateCreate
nvjpegEncoderStateDestroy
nvjpegEncodeYUV,(nvjpegHandle_t handle
nvjpegGetCudartProperty
nvjpegGetDeviceMemoryPadding
nvjpegGetImageInfo
nvjpegGetPinnedMemoryPadding
nvjpegGetProperty
nvjpegJpegStateCreate
nvjpegJpegStateDestroy
nvjpegJpegStreamCreate
nvjpegJpegStreamDestroy
nvjpegJpegStreamGetChromaSubsampling
nvjpegJpegStreamGetComponentDimensions
nvjpegJpegStreamGetComponentsNum
nvjpegJpegStreamGetFrameDimensions
nvjpegJpegStreamGetJpegEncoding
nvjpegJpegStreamParse
nvjpegJpegStreamParseHeader
nvjpegSetDeviceMemoryPadding
nvjpegSetPinnedMemoryPadding
nvjpegStateAttachDeviceBuffer
nvjpegStateAttachPinnedBuffer
       

For Linux x86_64, ARM and Power targets, Nsight Systems is capable of capturing information about the MPI APIs executed in the profiled process. It has built-in API trace support for Open MPI and MPICH based MPI implementations.

Only a subset of the MPI API, including blocking and non-blocking point-to-point and collective communication, and file I/O operations, is traced. If you require more control over the list of traced APIs or if you are using a different MPI implementation, you can use the NVTX wrappers for MPI. If you set the environment variable LD_PRELOAD to the path of generated wrapper library, Nsight Systems will capture and report the MPI API trace information when NVTX tracing is enabled. Choose an NVTX domain name other than "MPI", since it is filtered out by Nsight Systems when MPI tracing is not enabled.

MPI Communication Parameters

Nsight Systems can get additional information about MPI communication parameters. Currently, the parameters are only visible in the mouseover tooltips or in the eventlog. This means that the data is only available via the GUI. Future versions of the tool will export this information into the SQLite data files for postrun analysis.

In order to fully interpret MPI communications, data for all ranks associated with a communication operation must be loaded into Nsight Systems.

Here is an example of MPI_COMM_WORLD data. This does not require any additional team data, since local rank is the same as global rank.

(Screenshot shows communication parameters for an MPI_Bcast call on rank 3)

When not all processes that are involved in an MPI communication are loaded into Nsight Systems the following information is available.

• Right-hand screenshot shows a reused communicator handle (last number increased).
• Encoding: MPI_COMM[*team size*]*global-group-root-rank*.*group-ID*

When all reports are loaded into Nsight Systems:

• World rank is shown in addition to group-local rank "(world rank X)"
• Encoding: MPI_COMM[*team size*]{rank0, rank1, ...}
• At most 8 ranks are shown (the numbers represent world ranks, the position in the list is the group-local rank)

MPI functions traced:

MPI_Init[_thread], MPI_Finalize
MPI_Send, MPI_{B,S,R}send, MPI_Recv, MPI_Mrecv
MPI_Sendrecv[_replace]

MPI_Barrier, MPI_Bcast
MPI_Scatter[v], MPI_Gather[v]
MPI_Allgather[v], MPI_Alltoall[{v,w}]
MPI_Allreduce, MPI_Reduce[_{scatter,scatter_block,local}]
MPI_Scan, MPI_Exscan

MPI_Isend, MPI_I{b,s,r}send, MPI_I[m]recv
MPI_{Send,Bsend,Ssend,Rsend,Recv}_init
MPI_Start[all]
MPI_Ibarrier, MPI_Ibcast
MPI_Iscatter[v], MPI_Igather[v]
MPI_Iallgather[v], MPI_Ialltoall[{v,w}]
MPI_Iallreduce, MPI_Ireduce[{scatter,scatter_block}]
MPI_I[ex]scan
MPI_Wait[{all,any,some}]

MPI_File_{open,close,delete,sync}
MPI_File_{read,write}[_{all,all_begin,all_end}]
MPI_File_{read,write}_at[_{all,all_begin,all_end}]
MPI_File_{read,write}_shared
MPI_File_{read,write}_ordered[_{begin,end}]
MPI_File_i{read,write}[_{all,at,at_all,shared}]
MPI_File_set_{size,view,info}
MPI_File_get_{size,view,info,group,amode}
MPI_File_preallocate

If OpenSHMEM library trace is selected Nsight Systems will trace the subset of OpenSHMEM API functions that are most likely be involved in performance bottlenecks. To keep overhead low Nsight Systems does not trace all functions.

OpenSHMEM 1.5 Functions Not Traced

shmem_my_pe
shmem_n_pes
shmem_global_exit
shmem_pe_accessible
shmem_addr_accessible
shmem_ctx_{create,destroy,get_team}
shmem_global_exit
shmem_info_get_{version,name}
shmem_{my_pe,n_pes,pe_accessible,ptr}
shmem_query_thread
shmem_team_{create_ctx,destroy}
shmem_team_get_config
shmem_team_{my_pe,n_pes,translate_pe}
shmem_team_split_{2d,strided}
shmem_test*

If UCX library trace is selected Nsight Systems will trace the subset of functions of the UCX protocol layer UCP that are most likely be involved in performance bottlenecks. To keep overhead low Nsight Systems does not trace all functions.

UCX functions traced:

ucp_am_send_nb[x]
ucp_am_recv_data_nbx
ucp_am_data_release
ucp_atomic_{add{32,64},cswap{32,64},fadd{32,64},swap{32,64}}
ucp_atomic_{post,fetch_nb,op_nbx}
ucp_cleanup
ucp_config_{modify,read,release}
ucp_disconnect_nb
ucp_dt_{create_generic,destroy}
ucp_ep_{create,destroy,modify_nb,close_nbx}
ucp_ep_flush[{_nb,_nbx}]
ucp_listener_{create,destroy,query,reject}
ucp_mem_{advise,map,unmap,query}
ucp_{put,get}[_nbi]
ucp_{put,get}_nb[x]
ucp_request_{alloc,cancel,is_completed}
ucp_rkey_{buffer_release,destroy,pack,ptr}
ucp_stream_data_release
ucp_stream_recv_data_nb
ucp_stream_{send,recv}_nb[x]
ucp_stream_worker_poll
ucp_tag_msg_recv_nb[x]
ucp_tag_{send,recv}_nbr
ucp_tag_{send,recv}_nb[x]
ucp_tag_send_sync_nb[x]
ucp_worker_{create,destroy,get_address,get_efd,arm,fence,wait,signal,wait_mem}
ucp_worker_flush[{_nb,_nbx}]
ucp_worker_set_am_{handler,recv_handler}
      

UCX Functions Not Traced:

ucp_config_print
ucp_conn_request_query
ucp_context_{query,print_info}
ucp_get_version[_string]
ucp_ep_{close_nb,print_info,query,rkey_unpack}
ucp_mem_print_info
ucp_request_{check_status,free,query,release,test}
ucp_stream_recv_request_test
ucp_tag_probe_nb
ucp_tag_recv_request_test
ucp_worker_{address_query,print_info,progress,query,release_address}
       

Additional API functions from other UCX layers may be added in a future version of the product.

The NVIDIA network communication libraries NVSHMEM and NCCL have been instrumented using NVTX annotations. To enable tracing these libraries in Nsight Systems, turn on NVTX tracing in the GUI or CLI. To enable the NVTX instrumentation of the NVSHMEM library, make sure that the environment variable NVSHMEM_NVTX is set properly, e.g. NVSHMEM_NVTX=common.

This feature is experimental.

Overview

NVIDIA ConnectX smart network interface cards (smart NICs) offer advanced hardware offloads and accelerations for network operations. Viewing smart NICs metrics, on Nsight Systems timeline, enables developers to better understand their application’s network usage. Developers can use this information to optimize the application’s performance.

Limitations/Requirements

• NIC metric sampling supports NVIDIA ConnectX boards starting with ConnectX 5
• NIC metric sampling supported on Linux x86_64 machines only, having minimum Linux kernel 4.12 and minimum MLNX_OFED 4.1.
• NIC metric sampling is only available from the command line

Collecting NIC Metrics Using the Command Line

To collect NIC performance metric, using Nsight Systems CLI, add the --nic-metrics command line switch:

nsys profile --nic-metrics=true my_app

Available Metrics

• Bytes sent - Number of bytes sent through all NIC ports.
• Packets sent - Number of packets sent through all NIC ports.
• Bytes received - Number of bytes received by all NIC ports.
• Packets received - Number of packets received by all NIC ports.
• CNPs sent - Number of congestion notification packets sent by the NIC.
• CNPs received - Number of congestion notification packets received and handled by the NIC.
• Send waits - The number of ticks during which ports had data to transmit but no data was sent during the entire tick (either because of insufficient credits or because of lack of arbitration)

Usage Examples

• The Bytes sent/sec and the Bytes received/sec metrics enables identifying idle and busy NIC times.
  • Developers may shift network operations from busy to idle times to reduce network congestion and latency.
  • Developers can use idle NIC times to send additional data without reducing application performance.
• CNPs (congestion notification packets) received/sent and Send waits metrics may explain network latencies. A developer seeing the time periods when the network was congested may rewrite his algorithm to avoid the observed congestions.
• Comparing the ratio between bytes sent and packets sent can give developers a hint about the average packet size their application is using. The larger the packet size, the higher the practical bandwidth.

Users can generate a new report by stopping a profiling session. If a profiling session has been canceled, a report will not be generated, and all collected data will be discarded.

A new .nsys-rep file will be created and put into the same directory as the project file (.qdproj).

An existing .nsys-rep file can be opened using File > Open....

Report files (.nsys-rep) are self-contained and can be shared with other users of Nsight Systems. The only requirement is that the same or newer version of Nsight Systems is always used to open report files.

Project files (.qdproj) are currently not shareable, since they contain full paths to the report files.

To quickly navigate to the directory containing the report file, right click on it in the Project Explorer, and choose Show in folder... in the context menu.

While generating a new report or loading an existing one, a new tab will be created. The most important parts of the report tab are:

• View selector — Allows switching between Analysis Summary, Timeline View, Diagnostics Summary, and Symbol Resolution Logs views.

  

• Timeline — This is where all charts are displayed.

• Function table — Located below the timeline, it displays statistical information about functions in the target application in multiple ways.

Additionally, the following controls are available:

• Zoom slider — Allows you to vertically zoom the charts on the timeline.

This view shows a summary of the profiling session. In particular, it is useful to review the project configuration used to generate this report. Information from this view can be selected and copied using the mouse cursor.

The timeline view consists of two main controls: the timeline at the top, and a bottom pane that contains the events view and the function table. In some cases, when sampling of a process has not been enabled, the function table might be empty and hidden.

The bottom view selector sets the view that is displayed in the bottom pane.

This view shows important messages. Some of them were generated during the profiling session, while some were added while processing and analyzing data in the report. Messages can be one of the following types:

• Informational messages

• Warnings

• Errors

To draw attention to important diagnostics messages, a summary line is displayed on the timeline view in the top right corner:

Information from this view can be selected and copied using the mouse cursor.

This view shows all messages related to the process of resolving symbols. It might be useful to debug issues when some of the symbol names in the symbols table of the timeline view are unresolved.

When multiple reports are loaded into a single timeline, timestamps between them need to be adjusted, such that events that happened at the same time appear to be aligned.

Nsight Systems can automatically adjust timestamps based on UTC time recorded around the collection start time. This method is used by default when other more precise methods are not available. This time can be seen as UTC time at t=0 in the Analysis Summary page of the report file. Refer to your OS documentation to learn how to sync the software clock using the Network Time Protocol (NTP). NTP-based time synchronization is not very precise, with the typical errors on the scale of one to tens of milliseconds.

Reports collected on the same physical machine can use synchronization based on Timestamp Counter (TSC) values. These are platform-specific counters, typically accessed in user space applications using the RDTSC instruction on x86_64 architecture, or by reading the CNTVCT register on Arm64. Their values converted to nanoseconds can be seen as TSC value at t=0 in the Analysis Summary page of the report file. Reports synchronized using TSC values can be aligned with nanoseconds-level precision.

TSC-based time synchronization is activated automatically, when Nsight Systems detects that reports come from same target and that the same TSC value corresponds to very close UTC times. Targets are considered to be the same when either explicitly set environment variables NSYS_HW_ID are the same for both reports or when target hostnames are the same and NSYS_HW_ID is not set for either target. The difference between UTC and TSC time offsets must be below 1 second to choose TSC-based time synchronization.

To find out which synchronization method was used, navigate to the Analysis Summary tab of an added report and check the Report alignment source property of a target. Note, that the first report won’t have this parameter.

When loading multiple reports into a single timeline, it is always advisable to first check that time synchronization looks correct, by zooming into synchronization or communication events that are expected to be aligned.

When reports are added to the same timeline Nsight Systems will automatically line them up by timestamps as described above. If you want Nsight Systems to also recognize matching process or hardware information, you will need to set environment variables NSYS_SYSTEM_ID and NSYS_HW_ID as shown below at the time of report collection (such as when using "nsys profile ..." command).

When loading a pair of given report files into the same timeline, they will be merged in one of the following configurations:

• Different hardware — is used when reports are coming from different physical machines, and no hardware resources are shared in these reports. This mode is used when neither NSYS_HW_ID or NSYS_SYSTEM_ID is set and target hostnames are different or absent, and can be additionally signalled by specifying different NSYS_HW_ID values.

• Different systems, same hardware — is used when reports are collected on different virtual machines (VMs) or containers on the same physical machine. To activate this mode, specify the same value of NSYS_HW_ID when collecting the reports.

• Same system — is used when reports are collected within the same operating system (or container) environment. In this mode a process identifier (PID) 100 will refer to the same process in both reports. To manually activate this mode, specify the same value of NSYS_SYSTEM_ID when collecting the reports. This mode is automatically selected when target hostnames are the same and neither NSYS_HW_ID or NSYS_SYSTEM_ID is provided.

The following diagrams demonstrate typical cases:

A typical scenario is when a computing job is run using one of the MPI implementations. Each instance of the app can be profiled separately, resulting in multiple report files. For example:

# Run MPI job without the profiler:
mpirun <mpirun-options> ./myApp
# Run MPI job and profile each instance of the application:
mpirun <mpirun-options> nsys profile -o report-%p <nsys-options>./myApp

When each MPI rank runs on a different node, the command above works fine, since the default pairing mode (different hardware) will be used.

When all MPI ranks run the localhost only, use this command (value "A" was chosen arbitrarily, it can be any non-empty string):

NSYS_SYSTEM_ID=A mpirun <mpirun-options> nsys profile -o report-%p <nsys-options> ./myApp

For convenience, the MPI rank can be encoded into the report filename. For Open MPI, use the following command to create report files based on the global rank value:

mpirun <mpirun-options> nsys profile -o report-%q{OMPI_COMM_WORLD_RANK} <nsys-options> ./myApp

MPICH-based implementations set the environment variable PMI_RANK and Slurm (srun) provides the global MPI rank in SLURM_PROCID.

• Only report files collected with Nsight Systems version 2021.3 and newer are fully supported.

• Sequential reports collected in a single CLI profiling session cannot be loaded into a single timeline yet.

usage:

nsys [global-options] analyze [options]
       [nsys-rep-or-sqlite-file]

If a .nsys-rep file is given as the input file and there is no .sqlite file with the same name in the same directory, it will be generated.

Note: The Expert System view in the GUI will give you the equivalent command line.

The Expert System View can be found in the same drop-down as the Events View. If there is no .sqlite file with the same name as the .nsys-rep file in the same directory, it will be generated.

The highlighting is not supported for rules that do not return an event but rather an arbitrary time range (e.g. GPU utilization rules).

The CLI and GUI share the same rule scripts and messages. There might be some formatting differences between the output table in GUI and CLI.

Rules are scripts that run on the SQLite DB output from Nsight Systems to find common improvable usage patterns.

Each rule has an advice summary with explanation of the problem found and suggestions to address it. Only the top 50 results are displayed by default.

There are currently six rules in the expert system. They are described below. Additional rules will be made available in a future version of Nsight Systems.

Info

To find out report's start and end time use info command.

Usage:

ImportNvtxt --cmd info -i [--input] arg

Example:

ImportNvtxt info Report.nsys-rep
Analysis start (ns)	83501026500000
Analysis end (ns)	83506375000000

Create

You can create a report file using existing NVTXT with create command.

Usage:

ImportNvtxt --cmd create -n [--nvtxt] arg -o [--output] arg [-m [--mode] mode_name mode_args]

Usage for lerp mode is:

--mode lerp --ns_a arg --ns_b arg [--nvtxt_a arg --nvtxt_b arg]

If nvtxt_a and nvtxt_b are not specified, they are respectively set to nvtxt file's minimum and maximum time value.

Usage for lin mode is:

--mode lin --ns_a arg --freq arg [--nvtxt_a arg]

If nvtxt_a is not specified, it is set to nvtxt file's minimum time value.

Examples:

ImportNvtxt --cmd create -n Sample.nvtxt -o Report.nsys-rep

The output will be a new generated report file which can be opened and viewed by Nsight Systems.

Merge

To merge NVTXT file with an existing report file use merge command.

Usage:

ImportNvtxt --cmd merge -i [--input] arg -n [--nvtxt] arg -o [--output] arg [-m [--mode] mode_name mode_args]

Usage for lerp mode is:

--mode lerp --ns_a arg --ns_b arg [--nvtxt_a arg --nvtxt_b arg]

If nvtxt_a and nvtxt_b are not specified, they are respectively set to nvtxt file's minimum and maximum time value.

Usage for lin mode is:

--mode lin  --ns_a arg --freq arg [--nvtxt_a arg]

If nvtxt_a is not specified, it is set to nvtxt file's minimum time value.

Time values in <filename.nvtxt> are assumed to be nanoseconds if no mode specified.

Example

ImportNvtxt --cmd merge -i Report.nsys-rep -n Sample.nvtxt -o NewReport.nsys-rep

Profiling

Nsight Systems uses a settings file (NVIDIA Nsight Systems.ini) on the host to store information about loaded projects, report files, window layout configuration, etc. Location of the settings file is described in the Help → About dialog. Deleting the settings file will restore Nsight Systems to a fresh state, but all projects and reports will disappear from the Project Explorer.

Environment Variables

TMPDIR=/testdata ./bin/nsys profile -t cuda matrixMul
       

Environment variable control support for Windows target trace is not available, but there is a quick workaround:

• Create a batch file that sets the env vars and launches your application.
• Set Nsight Systems to launch the batch file as its target, i.e. set the project settings target path to the path of batch file.
• Start the trace. Nsight Systems will launch the batch file in a new cmd instance and trace any child process it launches. In fact, it will trace the whole process tree whose root is the cmd running your batch file.

WebGL Testing

“C:\Program Files (x86)\Google\Chrome\Application\chrome.exe”
       --inprocess-gpu --no-sandbox --disable-gpu-watchdog --use-angle=gl
       https://webglsamples.org/aquarium/aquarium.html
       

Common Issues with QNX Targets

• Make sure that tracelogger utility is available and can be run on the target.

• Make sure that /tmp directory is accessible and supports sub-directories.

• When switching between Nsight Systems versions, processes related to the previous version, including profiled applications forked by the daemon, must be killed before the new version is used. If you experience issues after switching between Nsight Systems versions, try rebooting the target.

If you have collected a report file using the CLI and the report will not open in the GUI, check to see that your GUI version is the same or greater than the CLI version you used. If it is not, download a new version of the Nsight Systems GUI and you will be able to load and visualize your report.

This situation occurs most frequently when you update Nsight Systems using a CLI only package, such as the package available from the NVIDIA HPC SDK.

In many cases, it is important to profile an application from the very beginning of its execution. When launching processes, Nsight Systems takes care of it by making sure that the profiling session is fully initialized before making the exec() system call on Linux.

If the process launch capabilities of Nsight Systems are not sufficient, the application should be launched manually, and the profiler should be configured to attach to the already launched process. One approach would be to call sleep() somewhere early in the application code, which would provide time for the user to attach to the process in Nsight Systems Embedded Platforms Edition, but there are two other more convenient mechanisms that can be used on Linux, without the need to recompile the application. (Note that the rest of this section is only applicable to Linux-based target devices.)

Both mechanisms ensure that between the time the process is created (and therefore its PID is known) and the time any of the application's code is called, the process is stopped and waits for a signal to be delivered before continuing.

If opening the Nsight Systems Linux GUI fails with one of the following errors, you may be missing some required libraries:

This application failed to start because it could not find or load the Qt platform plugin "xcb" in "". Available platform plugins are: xcb. Reinstalling the application may fix this problem.

or

error while loading shared libraries: [library_name]: cannot open shared object file: No such file or directory

If stack trace information is missing symbols and you have a symbol file, you can manually re-resolve using the ResolveSymbols utility. This can be done by right-clicking the report file in the Project Explorer window and selecting "Resolve Symbols...".

Alternatively, you can find the utility as a separate executable in the [installation_path]\Host directory. This utility works with ELF format files, with Windows PDB directories and symbol servers, or with files where each line is in the format <start><length><name>.

host-linux-x64/nvlog.config.template

When reporting any bugs please include the build version number as described in the Help → About dialog. If possible, attach log files and report (.nsys-rep) files, as they already contain necessary version information.

NVIDIA Deep Learning Institute Training - Self-Paced Online Course Optimizing CUDA Machine Learning Codes With Nsight Profiling Tools

2018 NCSA Blue Waters Webinar - Video Only Introduction to NVIDIA Nsight Systems

NVIDIA developer blogs, these are longer form, technical pieces written by tool and domain experts.

Short videos, only a minute or two, to introduce new features.

To file a bug report or to ask a question on the Nsight Systems forums, you will need to register with the NVIDIA Developer Program. See the FAQ. You do not need to register to read the forums.

After that, you can access Nsight Systems Forums and the NVIDIA Bug Tracking System.

To submit feedback directly from the GUI, go to Help->Send Feedback and fill out the form. Enter your email address if you would like to hear back from the Nsight Systems team.