NVlabs
NVlabs

cuda-oxide

cuda-oxide is an experimental Rust-to-CUDA compiler that lets you write (SIMT) GPU kernels in safe(ish), idiomatic Rust. It compiles standard Rust code directly to PTX — no DSLs, no foreign language bindings, just Rust.

Analisi AI · Italiano

openai · gpt-4o-mini

Sintesi

cuda-oxide è un compilatore sperimentale che consente di scrivere kernel GPU SIMT in Rust idiomatico e sicuro, compilando direttamente il codice Rust in PTX CUDA senza bisogno di DSL o binding a linguaggi esterni. Include un backend rustc personalizzato, runtime host per la gestione della memoria e il lancio dei kernel, e supporta funzioni generiche, closure con catture, e molte astrazioni lato device come memoria condivisa, barriere e operazioni warp/cluster. Il progetto è in fase alpha e mira a integrare la programmazione CUDA completa nativa in Rust con pipeline di compilazione avanzate. Fornisce anche supporto asincrono per la programmazione GPU con operazioni lazy e sincronizzazione esplicita.

Casi d'uso

  • Sviluppo di kernel GPU ad alte prestazioni utilizzando il linguaggio Rust senza bisogno di codice esterno o DSL specifici.
  • Progettazione di applicazioni di calcolo intensivo su GPU con uso di closure, funzioni generiche e sicurezza di tipo in Rust.
  • Realizzazione di pipeline asincrone di elaborazione GPU per machine learning o simulazioni scientifiche usando l'API asincrona nativa.
  • Sviluppo di librerie Rust per accelerazione GPU con interoperabilità tra moduli e supporto per device FFI con C++.
  • Creazione di strumenti di compilazione o di debugging per codice Rust GPU tramite la pipeline di compilazione personalizzata.

Idee SaaS / Business

Piattaforma Cloud per Compilazione Rust-to-GPU

Offrire un servizio cloud che consenta agli sviluppatori di compilare e testare kernel Rust su GPU senza installare toolchain locali, semplificando l'accesso a hardware NVIDIA e alla toolchain CUDA-oxide.

Tool di Ottimizzazione e Analisi Kernel GPU Rust

Fornire un SaaS che esegua analisi automatica del codice Rust per GPU, suggerendo ottimizzazioni a livello di compilazione, runtime e configurazioni di lancio, con report interattivi e suggerimenti.

Ambiente di Sviluppo As-a-Service per Rust CUDA

Offrire un IDE online integrato con supporto completo per cuda-oxide, inclusi debugging remoto su GPU, esecuzioni asincrone, gestione di progetti Rust-GPU e collaborazione real-time.

Target utenti
Sviluppatori Rust interessati alla programmazione GPU ad alte prestazioni, ricercatori e ingegneri che desiderano sfruttare l'ecosistema Rust per CUDA senza scrivere codice in C++ o DSL esterni, e comunità accademiche o industriali che sviluppano algoritmi paralleli ottimizzati per GPU NVIDIA.
Categoria
Rust, CUDA PTX, LLVM 21+ con backend NVPTX
Monetizzazione
Offrire servizi di supporto professionale, consulenza su integrazione e ottimizzazione CUDA-Rust, formazione specializzata, e piattaforme cloud a pagamento per compilazione e testing GPU; sviluppare funzionalità enterprise advanced da distribuire sotto licenza commerciale.
Licenza
Apache License 2.0
Trend: La crescente attenzione verso Rust e la programmazione GPU rende cuda-oxide un progetto innovativo, con forte interesse nella comunità di sviluppo parallelo e HPC.

README · tradotto in italiano

Introduzione

cuda-oxide è un backend rustc personalizzato per compilare kernel GPU in puro Rust. Il workspace combina:

  • Compilazione single-source — codice host e device convivono nello stesso file, compilabili con un unico comando cargo oxide build
  • Un backend di codegen rustc che compila funzioni #[kernel] in CUDA PTX
  • Astrazioni lato device (indicizzazione type-safe, memoria condivisa, atomici con ambito, barriere, TMA, operazioni warp/cluster)
  • Un runtime lato host per la gestione della memoria e il lancio dei kernel (cuda-core, cuda-async)
  • Una pipeline di compilazione Rust-native via Pliron, un framework IR simile a MLIR in Rust (Rust → Rust MIR → Pliron IR → LLVM IR → PTX)

Stato del Progetto

cuda-oxide è un compilatore sperimentale che mostra come scrivere kernel CUDA SIMT nativamente in Rust puro — senza DSL o binding esterni — e rendere la tecnologia accessibile alla comunità Rust. Il progetto è in fase iniziale (alpha), sotto sviluppo attivo: sono possibili bug, funzionalità incomplete e rotture API. Si incoraggia l'uso e il feedback degli utenti per guidarne l'evoluzione.

Per contribuire, consultare CONTRIBUTING.md.

Avvio Rapido

use cuda_device::{cuda_module, kernel, thread, DisjointSlice};
use cuda_core::{CudaContext, DeviceBuffer, LaunchConfig};

#[cuda_module]
mod kernels {
    use super::*;

    #[kernel]
    pub fn map<T: Copy, F: Fn(T) -> T + Copy>(f: F, input: &[T], mut out: DisjointSlice<T>) {
        let idx = thread::index_1d();
        let i = idx.get();
        if let Some(out_elem) = out.get_mut(idx) {
            *out_elem = f(input[i]);
        }
    }
}

fn main() {
    let ctx = CudaContext::new(0).unwrap();
    let stream = ctx.default_stream();

    let data: Vec<f32> = (0..1024).map(|i| i as f32).collect();
    let input = DeviceBuffer::from_host(&stream, &data).unwrap();
    let mut output = DeviceBuffer::<f32>::zeroed(&stream, 1024).unwrap();

    let module = kernels::load(&ctx).unwrap();

    let factor = 2.5f32;
    module
        .map::<f32, _>(
            &stream,
            LaunchConfig::for_num_elems(1024),
            move |x: f32| x * factor,
            &input,
            &mut output,
        )
        .unwrap();

    let result = output.to_host_vec(&stream).unwrap();
    assert!((result[1] - 2.5).abs() < 1e-5);
}

Questo esempio definisce un kernel generico #[kernel] che accetta qualsiasi chiusura Fn(T) -> T. L'attributo #[cuda_module] include l'artifact generato nel binario host e crea un metodo di lancio tipizzato module.map::<f32, _>(...). La closure move |x| x * factor viene catturata, scalarizzata e passata automaticamente ai parametri del kernel.

Per lavoro asincrono composabile su GPU, i parametri di stream scompaiono, la funzione map_async restituisce un oggetto lazy DeviceOperation e l'esecuzione avviene chiamando .sync() o .await.

use cuda_async::device_operation::DeviceOperation;

let factor = 2.5f32;
module
    .map_async::<f32, _>(
        LaunchConfig::for_num_elems(1024),
        move |x: f32| x * factor,
        &input,
        &mut output,
    )?
    .sync()?;
// oppure: .await?;

Setup

Requisiti

  • cargo-oxide — comando cargo per gestire la build pipeline (cargo oxide run, etc.)
  • Rust nightly con componenti rust-src e rustc-dev (con toolchain bloccata in rust-toolchain.toml)
  • CUDA Toolkit (versione 12.x o superiore)
  • LLVM 21+ con backend NVPTX (llc accessibile in PATH)
  • Clang + header di sviluppo per bindgen nel crate host cuda-bindings
  • Linux (testato su Ubuntu 24.04)

Note su LLVM 21

Usiamo intrinsic TMA e simili supportati solo da llc >= 21. Kernel semplici possono funzionare su versioni LLVM inferiori, ma funzioni Blackwell+ richiedono la versione 21 o superiore.

Installazione

cargo-oxide

Nel repo cuda-oxide, cargo oxide è disponibile come workspace alias. Per usarlo in altri progetti:

cargo install --git https://github.com/NVlabs/cuda-oxide.git cargo-oxide

Al primo utilizzo scaricherà e compilerà automaticamente il backend.

Rust

rustup toolchain install nightly-2026-04-03
rustup component add rust-src rustc-dev --toolchain nightly-2026-04-03

CUDA

export PATH="/usr/local/cuda/bin:$PATH"
nvcc --version

LLVM

sudo apt install llvm-21

Se non disponibile, usare lo script di llvm.org:

sudo apt-get install -y lsb-release wget software-properties-common gnupg
wget https://apt.llvm.org/llvm.sh && chmod +x llvm.sh
sudo ./llvm.sh 21

Verificare supporto NVPTX:

llc-21 --version | grep nvptx

Clang

Per bindgen, installare clang completo:

sudo apt install clang-21

cargo oxide doctor verificherà la corretta configurazione.

Container di sviluppo

È disponibile un devcontainer standard in .devcontainer/ per ambiente riproducibile CUDA, LLVM, Clang e Rust.

Verifica installazione

cargo oxide doctor
cargo oxide run vecadd

Il comando run vecadd compila e avvia un kernel Rust su GPU, stampando ✓ SUCCESS: All 1024 elements correct! in caso di successo.

Esempi

Sono disponibili 46 esempi in crates/rustc-codegen-cuda/examples/. Alcuni esempi principali:

EsempioDescrizione
vecaddAddizione vettoriale — esempio canonico di partenza
host_closureKernel generici con closure passate da host
genericKernel generici con monomorfismo
gemm_solGEMM SoL: 868 TFLOPS (58% cuBLAS su B200), 8 kernel in 4 fasi
tcgen05Tensor cores Blackwell (sm_100a): TMEM, MMA, gruppi CTA
atomicsAtomici GPU: test per 6 tipi, diversi ambiti e ordinamenti
clusterCluster di thread block + scambio a anello di DSMEM (Hopper+)
async_mlpPipeline asincrona MLP: GEMM → MatVec → ReLU su flussi concorrenti
mathdx_ffi_testFFT thread-level cuFFTDx + GEMM block-level cuBLASDx
async_vecaddEsecuzione GPU asincrona con cuda-async e DeviceOperation
cross_crate_kernelKernel definiti in crate libreria combinati in binari

Esempi d'uso:

cargo oxide run vecadd
cargo oxide run gemm_sol

Panoramica crate

Crate lato utente

CrateDescrizione
cuda-deviceIntrinsic device (thread::*, warp::*, barriere)
cuda-hostCaricamento moduli tipizzati, helper di lancio, LTOIR loader
cuda-macrosMacro procedurali (#[cuda_module], #[kernel], gpu_printf!)
cuda-bindingsBindings raw FFI a cuda.h
cuda-coreWrapper RAII sicuri (CudaContext, CudaStream, DeviceBuffer<T>)
cuda-asyncLivello esecuzione asincrona (DeviceOperation, DeviceFuture)
libnvvm-sysBindings a libNVVM per cuda-host::ltoir
nvjitlink-sysBindings a nvJitLink per cuda-host::ltoir

Compiler

CrateDescrizione
rustc-codegen-cudaBackend rustc personalizzato
mir-importerTraduzione Rust MIR → dialect-mir + pipeline
mir-lowerLowering dialect-mirdialect-llvm
dialect-mirDialetto pliron per Rust MIR
dialect-llvmDialetto pliron per LLVM IR
dialect-nvvmDialetto pliron per NVVM intrinsics

Build Tooling

CrateDescrizione
cargo-oxideSottocomando cargo (cargo oxide run etc.)

Documentazione

DirectoryDescrizione
cuda-oxide-bookDocumentazione principale: guida, internals, API reference

Stato Caratteristiche

  • Compilazione completa da Rust a PTX
  • Compilazione single-source host + device
  • Supporto per funzioni generiche con monomorfizzazione
  • Closure con catture tramite HMM
  • Strutture, enum e pattern matching lato device
  • Ampio supporto agli intrinsic GPU (thread, warp, memoria condivisa, barriere, TMA, cluster, atomici)
  • Kernel cross-crate
  • Generatione LTOIR per Blackwell+ (LTO device side)
  • FFI device Rust ↔ C++/CCCL via LTOIR
  • Integrazione MathDx: cuFFTDx FFT thread-level, cuBLASDx GEMM block-level
  • Runtime host sincronico (cuda-core) e asincrono (cuda-async)
  • GEMM SoL raggiunge 868 TFLOPS (58% di cuBLAS SoL) su B200

Documentazione

La cuda-oxide book è la guida di riferimento principale, con dettagli su creazione kernel SIMT in Rust, programmazione GPU sincrona e asincrona, architettura del compilatore e altro.

Per compilare e servire la documentazione localmente, vedere cuda-oxide-book/README.md.

Ecosistema

cuda-oxide è uno dei vari progetti Rust+GPU in sviluppo attivo. Altri affrontano ambiti diversi come Vulkan/SPIR-V per la grafica, offload implicito con LLVM, backend CUDA terze parti, binding driver sicuri. Il progetto collabora con la comunità Rust GPU per far avanzare la programmazione GPU in Rust in modo integrato.

Licenza

Il crate cuda-bindings è sotto NVIDIA Software License: LICENSE-NVIDIA. Gli altri crate sono sotto Apache License, Versione 2.0: LICENSE-APACHE.

Attività commit · ultime 26 settimane

stima

Approfondimenti AI

L'AI sta preparando gli approfondimenti…

Chiedi al repo

AI · contesto README + issue

Fai una domanda sul progetto. L'AI legge README e issue recenti.

Sponsor · Sconto esclusivo RepoRadar AI

Hai bisogno di un server per far girare NVlabs/cuda-oxide?

Abbiamo testato decine di provider e Hostinger VPS è il miglior rapporto qualità/prezzo per self-hostare le repo che trovi qui. Setup in 1 click, pannello semplice e supporto 24/7.

Deploy in 1 click
2 vCPU · 8 GB RAM · NVMe
Backup + DDoS inclusi
Attiva sconto Hostinger VPSLink affiliato — supporti RepoRadar senza costi extra per te.

Questo progetto esiste grazie a voi

RepoRadar AI è gratis e senza pubblicità. Le donazioni coprono server, API e modelli AI.

Ogni analisi tradotta che leggi costa qualche centesimo di chiamate al modello. Se RepoRadar ti ha fatto risparmiare tempo, considera una piccola donazione cripto — anche pochi euro aiutano a mantenere il servizio libero per tutti.

Ethereum
ETH
0x86ECDF546d8dFc0739d44c066A6110F11cdB7773
Bitcoin
BTC
bc1qqe0wcmhnt78enk8ql0lxvey4z8hquxsxjtyz8r
Solana
SOL
EtTK61Lz7kfdDM8543TMMiAUUTbFVpzX5tvPEcBtZ3aj

Grazie di cuore — ogni contributo conta.