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-miniSintesi
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
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.
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.
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.
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-srcerustc-dev(con toolchain bloccata inrust-toolchain.toml) - CUDA Toolkit (versione 12.x o superiore)
- LLVM 21+ con backend NVPTX (
llcaccessibile in PATH) - Clang + header di sviluppo per
bindgennel crate hostcuda-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:
| Esempio | Descrizione |
|---|---|
vecadd | Addizione vettoriale — esempio canonico di partenza |
host_closure | Kernel generici con closure passate da host |
generic | Kernel generici con monomorfismo |
gemm_sol | GEMM SoL: 868 TFLOPS (58% cuBLAS su B200), 8 kernel in 4 fasi |
tcgen05 | Tensor cores Blackwell (sm_100a): TMEM, MMA, gruppi CTA |
atomics | Atomici GPU: test per 6 tipi, diversi ambiti e ordinamenti |
cluster | Cluster di thread block + scambio a anello di DSMEM (Hopper+) |
async_mlp | Pipeline asincrona MLP: GEMM → MatVec → ReLU su flussi concorrenti |
mathdx_ffi_test | FFT thread-level cuFFTDx + GEMM block-level cuBLASDx |
async_vecadd | Esecuzione GPU asincrona con cuda-async e DeviceOperation |
cross_crate_kernel | Kernel definiti in crate libreria combinati in binari |
Esempi d'uso:
cargo oxide run vecadd
cargo oxide run gemm_sol
Panoramica crate
Crate lato utente
| Crate | Descrizione |
|---|---|
cuda-device | Intrinsic device (thread::*, warp::*, barriere) |
cuda-host | Caricamento moduli tipizzati, helper di lancio, LTOIR loader |
cuda-macros | Macro procedurali (#[cuda_module], #[kernel], gpu_printf!) |
cuda-bindings | Bindings raw FFI a cuda.h |
cuda-core | Wrapper RAII sicuri (CudaContext, CudaStream, DeviceBuffer<T>) |
cuda-async | Livello esecuzione asincrona (DeviceOperation, DeviceFuture) |
libnvvm-sys | Bindings a libNVVM per cuda-host::ltoir |
nvjitlink-sys | Bindings a nvJitLink per cuda-host::ltoir |
Compiler
| Crate | Descrizione |
|---|---|
rustc-codegen-cuda | Backend rustc personalizzato |
mir-importer | Traduzione Rust MIR → dialect-mir + pipeline |
mir-lower | Lowering dialect-mir → dialect-llvm |
dialect-mir | Dialetto pliron per Rust MIR |
dialect-llvm | Dialetto pliron per LLVM IR |
dialect-nvvm | Dialetto pliron per NVVM intrinsics |
Build Tooling
| Crate | Descrizione |
|---|---|
cargo-oxide | Sottocomando cargo (cargo oxide run etc.) |
Documentazione
| Directory | Descrizione |
|---|---|
cuda-oxide-book | Documentazione 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
stimaApprofondimenti AI
Chiedi al repo
AI · contesto README + issueFai una domanda sul progetto. L'AI legge README e issue recenti.
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.
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.
0x86ECDF546d8dFc0739d44c066A6110F11cdB7773bc1qqe0wcmhnt78enk8ql0lxvey4z8hquxsxjtyz8rEtTK61Lz7kfdDM8543TMMiAUUTbFVpzX5tvPEcBtZ3ajGrazie di cuore — ogni contributo conta.