ZLUDA Documentation

Rust cuda & rust repository by @vosen. Model Context Protocol (MCP) server & tools with 15k stars.

14,724 stars Rust Rust 1 file · ~5,364 tokens #cuda#rust
RAW Doc

File: README.md

ZLUDA is a drop-in replacement for CUDA on non-NVIDIA GPUs. ZLUDA allows running unmodified CUDA applications using non-NVIDIA GPUs with near-native performance

File: detours-sys/README.md

Detours-sys

Rust bindings to Microsoft Detours

Usage

See the Detours Documentation for details.

Details

Use the buildtime_generation feature to generate bindings at build time,
otherwise use pregenerated bindings.

Note that this will require a relatively recent version of Clang be installed.

License

Licensed under either of

at your option.

Contribution

Unless you explicitly state otherwise, any contribution intentionally submitted
for inclusion in the work by you, as defined in the Apache-2.0 license, shall be
dual licensed as above, without any additional terms or conditions.

File: docs/src/building.md

Building

Dependencies

Build steps

  • Git clone the repo (make sure to use --recursive option to fetch submodules):
    git clone --recursive https://github.com/vosen/ZLUDA.git
  • Enter freshly cloned ZLUDA directory and build with cargo (this takes a while):
    • cargo xtask --release for Release build
    • cargo xtask for Debug build

File: docs/src/faq.md

FAQ

[!WARNING]
For legal reasons we can't help you with the pre-rollback versions (older than 4). See more here: https://www.theregister.com/2024/08/09/amd_zluda_take_down

General

  1. How I can donate to ZLUDA development?

    The ZLUDA project is fully funded and only accepts donations of labor.

  2. What organization is funding ZLUDA development?

    This will be revealed in due time.

  3. How can I follow ZLUDA's progress

Hardware

  1. AMD GPU support?

    ZLUDA supports AMD Radeon RX 5000 series and newer GPUs (both desktop and integrated).
    Older consumer GPUs (Polaris, Vega, etc.) and server‑class GPUs are not supported; these architectures differ significantly from recent desktop GPUs and would require substantial engineering effort.
    We expect that the near-future unified GPU architecture (UDNA) will be more similar to desktop GPUs.

  2. Intel GPU support?

    ZLUDA previously supported Intel GPUs, but not currently. It is possible to revive the Intel backend. The development team is focusing on high‑quality AMD GPU support and welcomes contributions.

  3. NVIDIA GPU support?

    Unlikely to ever be on the roadmap, because NVIDIA users can use the original CUDA. That said, if someone wants to add support we are open to contributions.

  4. Qualcomm GPU support?

    It would be interesting to have Qualcomm GPU support, but the development team is focusing on high‑quality AMD GPU support. We welcome contributions.

  5. macOS support?

    Unlikely to ever happen. There is very little non‑deprecated CUDA software for macOS, and what remains will soon be unsupported.

  6. ZLUDA on top of OpenCL or Vulkan?

    ZLUDA could be ported to OpenCL or Vulkan, but with significantly reduced functionality. This might be acceptable for a narrow use case, but it would not be as general‑purpose as using the native backend. Examples of features available with the current compilation path that are not exposed by either Vulkan or OpenCL:

    • Disabling FP contraction
    • Explicit alignment
    • Some subgroup and group operations
    • Bindless images
    • Pointer casts
    • Arbitrary virtual calls
    • Inline assembly
    • Rounding modes
    • Denormal modes

    Additionally, performance libraries (cuBLAS, cuDNN, etc.) cannot be easily mapped through Vulkan or OpenCL.

Software

  1. PyTorch support?

    PyTorch support is currently our top priority. We expect to have initial support fourth quarter of 2025.

  2. Tensorflow support?

    Tensorflow support is currently a top priority for ZLUDA and will follow PyTorch support.

  3. Blender support

    Blender is not on the roadmap, but it's often requested. Support might be added at certain point, but it's a low priority. If ZLUDA supports Blender, it will not support hardware ray-tracing (see Hardware ray-tracing (OptiX) support? section below).

  4. Hardware ray-tracing (OptiX) support?

    OptiX support is exceedingly complex. While it's built on top of CUDA, it uses its own dialect of PTX, uses its own host code and requires its own specific optimizations. It's unlikely that ZLUDA will ever support OptiX again. OptiX would require a very dedicated contributor (or team of contributors) to step in.

  5. Support for games using 32 bit PhysX?

    We are convinced that it's possible (both for AMD GPUs and NVIDIA GPUs). Necessary groundwork has been done (log collection) and there is a plan how to implement the feature. It's not on the roadmap and we are hoping for outside contributors to step in.

  6. Support for games using 64 bit PhysX (GameWorks)?

    It is definitely possible, pre-rollback ZLUDA had this capability. It's not on the roadmap and would require outside contributions.

  7. DLSS support?

    Previously DLSS support was blocked by a missing functionality in AMD's Direct3D driver: ability to enqueue HIP kernels into Direct3D command list. This functionality now ships in the newest driver and DLSS support should be possible. It's not on our roadmap, but if someone steps in to implement it, we'd be happy to merge.

File: docs/src/hip_sdk.md

Installing HIP SDK

On Windows, in addition to installing the GPU driver, you need to install the HIP SDK. Choose one of the two options below:

Official HIP SDK Nightly HIP SDK builds
✅ Automatic installation\
✅ Stable and supported by AMD\
❌ Older code\
❌ No Machine Learning support (PyTorch and TensorFlow will not work)

Installation

  1. Visit the AMD HIP SDK for Windows website
  2. Download the latest version for your OS
  3. Run the installer and follow the on-screen instructions | ❌ Manual installation
    ❌ Nightly builds by AMD - no stability guarantees
    ✅ Newer code
    ✅ Machine Learning support (required for PyTorch, TensorFlow, etc.)

Installation

  1. Visit the ROCm SDK nightly tarballs website
  2. Download a recent therock-dist-windows-gfx...tar.gz file:
    • If you don't know your , check the TechPowerUp GPU database. Example: for [AMD Radeon RX 9070](https://www.techpowerup.com/gpu-specs/radeon-rx-9070.c4250), the (Shader ISA) is 1201
    • Alternatively, download any tar.gz file, unpack it and run the included hipInfo.exe to determine your `` (gcnArchName)
  3. Extract the .tar.gz file to a directory of your choice (you can use 7-Zip)
    • Note: You may need to extract twice (first the .tar.gz, then the .tar file)
  4. Set the HIP_PATH environment variable to the extracted directory path (instructions here)
    • The directory must contain a bin subdirectory
    • The bin directory must contain various .dll files, including rocblas.dll |

Verification

ZLUDA includes a small CUDA application that tests the loading and initialization of all performance libraries. Run it using:

text
zluda.exe -- cuda_check.exe

The output should look like this:

text
nvcuda    : OK (C:\hip_sdk\bin\amdhip64_7.dll)
nvml      : OK
cufft11   : OK
cudnn9    : OK (C:\hip_sdk\bin\MIOpen.dll)
cudnn8    : OK (C:\hip_sdk\bin\MIOpen.dll)
cublaslt13: OK (C:\hip_sdk\bin\libhipblaslt.dll)
cusparse12: OK
cufft12   : OK
cublas13  : OK (C:\hip_sdk\bin\rocblas.dll)
cublaslt12: OK (C:\hip_sdk\bin\libhipblaslt.dll)
cublas12  : OK (C:\hip_sdk\bin\rocblas.dll)
cusparse11: OK

The path in parentheses is the path to the underlying HIP SDK library.

Caveats and Known Issues:

  • The path shown in parentheses indicates the underlying HIP SDK library, but is not guaranteed to be used. If an application loads the library from a different path before loading ZLUDA, ZLUDA will use the already-loaded library instead
  • cuda_check.exe may sometimes hang and not close due to a bug in MIOpen
  • When using the official HIP SDK, cuda_check.exe will fail to load cudnn8 and cudnn9 because the official SDK does not include MIOpen

File: docs/src/llama_cpp.md

llama.cpp

llama.cpp runs at native speed when compiled for CUDA architecture 86 and with cuBLAS enabled:

text
cmake -B build -DGGML_CUDA=ON -DCMAKE_CUDA_ARCHITECTURES="86" -DGGML_CUDA_FORCE_CUBLAS=true

Compiling for multiple CUDA architectures should be fine as long as one of the architectures is 80, 86 or 89.
Compiling with cuBLAS disabled might lead to performance degradation.

Windows

You need to install HIP SDK to have access to rocBLAS

File: docs/src/precompiling.md

Precompiling

Consider precompiling the GPU code with zluda_precompile if you are trying to run a large application.
zluda_precompile scans the location, extracts all the GPU code, compiles it, and saves it to the cache.
This way, the GPU code is already in the cache when the application is launched for the first time.

This process uses all the threads on the machine (making it faster than leaving it to the application) and may compile more code than is necessary for your application (making it potentially slower than leaving it to the application). Your mileage might vary.

Usage:

Windows:

text
zluda_precompile.exe

Linux:

text
zluda_precompile

where `` is the path to the directory or file.

File: docs/src/quick_start.md

Quick start

[!WARNING]
This version of ZLUDA is under heavy development and will likely not work with your application yet. In the meantime, you are encouraged to try it and report results.

How to get it

ZLUDA evolves quickly, we recommend downloading the most recent pre-release version.
Periodically, we mark a pre-release version as stable, you can download it from the "Releases" section on our Github page.

Usage

Windows

If you want to run PhysX check here. Instructions below apply to non-PhysX applications. You don't need HIP SDK for PhysX.

You should have a recent AMD GPU driver ("AMD Software: Adrenalin Edition") and HIP SDK installed.
To run your application either:

  • (Recommended) Use ZLUDA launcher:
    text
    <ZLUDA_DIRECTORY>\zluda.exe -- <APPLICATION> <APPLICATION_ARGUMENTS>
  • Copy all ZLUDA files (including nvcuda.dll) from zluda (if you downloaded a zip package) or target\release (if you built from sources) into a path which your application uses to load CUDA. Paths vary application to application, but usually it's the directory where the .exe file is located

Linux

Run your application like this:

  • Recommended method

    text
    LD_LIBRARY_PATH="<ZLUDA_DIRECTORY>:$LD_LIBRARY_PATH" <APPLICATION> <APPLICATION_ARGUMENTS>

    where <ZLUDA_DIRECTORY> is the directory which contains ZLUDA-provided libcuda.so: zluda if you downloaded a prebuilt package or target/release if you built from sources.

  • Alternative method

    text
    LD_AUDIT="<ZLUDA_DIRECTORY>/zluda_ld:$LD_AUDIT" <APPLICATION> <APPLICATION_ARGUMENTS>

    where <ZLUDA_DIRECTORY> is the directory which contains ZLUDA-provided libcuda.so: zluda if you downloaded a prebuilt package or target/release if you built from sources.

macOS

Not supported

File: docs/src/SUMMARY.md

Summary

General

Troubleshooting

Applications

For developers

File: docs/src/troubleshooting.md

Logging

Introduction

zluda_trace is a shim
for the CUDA API which traces application's CUDA usage. If your application is
encountering issues with ZLUDA, you should use zluda_trace to find out where and
why ZLUDA fails.

Quick Start

Linux

Run your application like this:

AMD GPU:

bash
ZLUDA_CUDA_LIB=<ZLUDA_DIRECTORY>/libcuda.so LD_LIBRARY_PATH=<ZLUDA_DIRECTORY>/trace/ \
ZLUDA_LOG_DIR=<LOG_DIRECTORY> <APPLICATION> <APPLICATION_ARGUMENTS>

NVIDIA GPU:

bash
LD_LIBRARY_PATH=<ZLUDA_DIRECTORY>/trace/ \
ZLUDA_LOG_DIR=<LOG_DIRECTORY> <APPLICATION> <APPLICATION_ARGUMENTS>

If you're filing a GitHub
issue, please create
an archive with your logs from <LOG_DIRECTORY> and attach it to the issue:

bash
tar -cvf logs.tar.gz -C <LOG_DIRECTORY> .

Windows

Run your application like this:

AMD GPU:

bash
zluda.exe --zluda-trace -- <APPLICATION> <APPLICATION_ARGUMENTS>

NVIDIA GPU:

bash
zluda.exe --nvidia-trace -- <APPLICATION> <APPLICATION_ARGUMENTS>

This will create a new directory with executable name with the log output in C:\Users\%USERNAME%\AppData\Local\Temp\zluda (%TEMP%\zluda).

If you're filing a GitHub
issue, please create
a .zip file with your logs from C:\Users\%USERNAME%\AppData\Local\Temp\zluda to attach to the
issue. In Windows Explorer, right click on the directory and select
"Send to"/"Compressed (zipped) folder". Exact steps may vary between
Windows versions.

Explanation

`LD_LIBRARY_PATH=/trace/`

<ZLUDA_DIRECTORY> is the directory that contains the ZLUDA driver (libcuda.so
and various other libraries). It will be target/release if you built from
source, or zluda if you downloaded one of the release packages. <ZLUDA_DIRECTORY>/trace
contains tracing shims for libcuda.so (zluda_trace) and other CUDA libraries.

[!NOTE]
LD_LIBRARY_PATH is an environment variable used by ld, Linux's dynamic linker. It tells ld that when it's looking for a shared library – for example, libcuda.so – it should first look in a specific list of directories, before system paths. It's just like PATH, but for shared libraries instead of executables.

`ZLUDA_CUDA_LIB=/libcuda.so`

By default, zluda_trace will log all calls and then redirect them to an actual
CUDA driver (libcuda.so). In order to use ZLUDA, ZLUDA_CUDA_LIB must be set
to the libcuda.so provided by ZLUDA. If ZLUDA_CUDA_LIB is not set,
zluda_trace will use NVIDIA’s libcuda.so.

`ZLUDA_LOG_DIR=`

By default, zluda_trace prints logs to stderr. In order to save them to a
file, as well as save other useful information, you must provide a
directory that they should be saved in – for example, /tmp/zluda.

Understanding the zluda_trace output

Let's look at the zluda_trace output for a simple application. Here's a
CUDA program that adds two numbers on the GPU:

cpp,linenos
#include 

__global__ void add(int a, int b, int *out) {
    *out = a + b;
}

int main() {
    int *result;
    cudaMallocManaged(&result, sizeof(int));
    add<<<1, 1>>>(1, 2, result);
    cudaDeviceSynchronize();
    std::cout << "result: " << *result << std::endl;
    cudaFree(result);
    return 0;
}

I've saved this file as add.cu. ZLUDA doesn't successfully run this
application yet, so I'll compile it and run it using zluda_trace and CUDA
in order to demonstrate all of zluda_trace's features.

bash
nvcc add.cu -o add -arch sm_80
LD_LIBRARY_PATH=~/ZLUDA/target/release/trace/ ZLUDA_LOG_DIR=/tmp/zluda ./add

The last few lines should look something like:

text
[ZLUDA_TRACE] cuCtxSynchronize() -> CUDA_SUCCESS
result: 3
[ZLUDA_TRACE] {CONTEXT_LOCAL_STORAGE_INTERFACE_V0301}::context_local_storage_get(value: 0x562c764a73c0, cu_ctx: 0x0, key: 0x562c764ba130) -> CUDA_SUCCESS
[ZLUDA_TRACE] cuMemFree_v2(dptr: 0x7f3ca2000000) -> CUDA_SUCCESS
[ZLUDA_TRACE] {CONTEXT_LOCAL_STORAGE_INTERFACE_V0301}::context_local_storage_delete(context: 0x562c764ba760, key: 0x562c764ba130) -> CUDA_ERROR_DEINITIALIZED
[ZLUDA_TRACE] cuLibraryUnload(library: 0x562c773ffb10) -> CUDA_ERROR_DEINITIALIZED
[ZLUDA_TRACE] cuDevicePrimaryCtxRelease(dev: 0) -> CUDA_ERROR_DEINITIALIZED

Now, let's take a look at our log directory:

bash
ls /tmp/zluda
add

zluda_trace creates a new directory for each run, based on the name of
the command. If the add directory already existed, it'd create an add_1
directory, and so on. Next, let's look at that newly-created directory:

bash
ls /tmp/zluda/add/
log.txt  module_0001_01.elf  module_0001_02.ptx

Let's take a look at each of these files.

log.txt

bash
#no_wrap
cat /tmp/zluda/add/log.txt
# ...
# cuModuleGetFunction(hfunc: 0x55ee94d645d0, hmod: 0x55ee94d63c40, name: "_Z3addiiPi") -> CUDA_SUCCESS
# cuLaunchKernel(f: 0x55ee94d645d0, gridDimX: 1, gridDimY: 1, gridDimZ: 1, blockDimX: 1, blockDimY: 1, blockDimZ: 1, sharedMemBytes: 0, hStream: 0x0, kernelParams: 0x7fffe0fa193c, extra: NULL) -> CUDA_SUCCESS
# {CONTEXT_LOCAL_STORAGE_INTERFACE_V0301}::context_local_storage_get(value: 0x55ee93e083c0, cu_ctx: 0x0, key: 0x55ee93e1b130) -> CUDA_SUCCESS
# cuCtxSynchronize() -> CUDA_SUCCESS
# {CONTEXT_LOCAL_STORAGE_INTERFACE_V0301}::context_local_storage_get(value: 0x55ee93e083c0, cu_ctx: 0x0, key: 0x55ee93e1b130) -> CUDA_SUCCESS
# cuMemFree_v2(dptr: 0x7fbde6000000) -> CUDA_SUCCESS
# {CONTEXT_LOCAL_STORAGE_INTERFACE_V0301}::context_local_storage_delete(context: 0x55ee93e1b760, key: 0x55ee93e1b130) -> CUDA_ERROR_DEINITIALIZED
# cuLibraryUnload(library: 0x55ee94d60ae0) -> CUDA_ERROR_DEINITIALIZED
# cuDevicePrimaryCtxRelease(dev: 0) -> CUDA_ERROR_DEINITIALIZED

As you can see, this is the same log that was written to stderr. It
records each call made to a CUDA library, the arguments it was passed,
and the status code returned. Most of these will be calls that you can
find in the NVIDIA documentation – for example,
cuModuleGetFunction
– but some of them aren't publicly documented.

For example, look at the calls to
{CONTEXT_LOCAL_STORAGE_INTERFACE_V0301}​::​context_local_storage_get.
Calls with this format are to what we call NVIDIA's Dark API. We'll
write more documentation for this later, but for now all you need to
know are that these are from function pointer tables returned by
cuGetExportTable.

We're looking at a very simple example, so it doesn't use any
performance libraries. If you use zluda_trace for code calling one of
NVIDIA's performance libraries, zluda_trace will log both the call to
that library, and then all of the calls made by that library call. That
looks like:

text
cublasCreate_v2(handle: 0x55e502373120) -> CUBLAS_STATUS_SUCCESS
    cuGetProcAddress_v2(symbol: "", pfn: 0x0, cudaVersion: 0, flags: 0, symbolStatus: NULL) -> CUDA_ERROR_NOT_FOUND

The call to cublasCreate_v2 is making a call to cuGetProcAddress_v2.

module_0001_01.elf

This is precompiled SASS assembly for a single GPU architecture.

module_0001_02.ptx

This is PTX assembly that is portable across many NVIDIA GPUs.

bash
cat /tmp/zluda/add/module_0001_02.ptx
# //
# //
# //
# //
# //
# //
#
# .version 8.7
# .target sm_80
# .address_size 64
#
# //
#
# .visible .entry _Z3addiiPi(
# .param .u32 _Z3addiiPi_param_0,
# .param .u32 _Z3addiiPi_param_1,
# .param .u64 _Z3addiiPi_param_2
# )
# {
# .reg .b32 %r<4>;
# .reg .b64 %rd<3>;
#
#
# ld.param.u32 %r1, [_Z3addiiPi_param_0];
# ld.param.u32 %r2, [_Z3addiiPi_param_1];
# ld.param.u64 %rd1, [_Z3addiiPi_param_2];
# cvta.to.global.u64 %rd2, %rd1;
# add.s32 %r3, %r2, %r1;
# st.global.u32 [%rd2], %r3;
# ret;
#
# }

This is the add function from add.cu. _Z3addiiPi is the add(int, int, int*) after C++ name mangling.

Compiler logs

There's one more kind of file zluda_trace might produce: a compiler error
log file. When zluda_trace encounters a PTX module, it tries to compile
it with ZLUDA's PTX compiler. Any errors produced will be saved into a
module_NNNN_NN.log file. For example, it might look like

text
Unrecognized statement "nanosleep.u32 %r101;"

We use this information to discover which PTX instructions are used by the
application and not supported by ZLUDA.

Explore Related Technologies & Topics

View All 9,111+ Repositories →

Discover similar high-velocity repositories, agent skills, and OpenAPI specifications across the ecosystem.

#cuda Browse cuda catalog → #rust Browse rust catalog → ⚡ MCP Servers Model Context Protocol → 🧠 Agent Skills Prompt & rule files → 🔌 OpenAPI Specs REST & Swagger schemas →