Update optimize-for-gpu skill documentation and assets
- Enhanced the skill description for clarity and accuracy regarding GPU acceleration. - Updated compatibility notes to reflect requirements for RAPIDS 26.06 and Python 3.11+. - Revised the library selection guidance to emphasize preferred usage patterns and legacy considerations. - Improved the optimization workflow section with detailed steps for defining contracts and checking suitability. - Added new code transformation patterns and clarified the use of cuSpatial and cuVS. - Updated the cuCIM reference to include installation instructions and performance characteristics. - Replaced the existing optimize-for-gpu.png image with a new version to better illustrate the skill's capabilities.
This commit is contained in:
Binary file not shown.
|
Before Width: | Height: | Size: 1.4 MiB After Width: | Height: | Size: 1.4 MiB |
@@ -1,14 +1,18 @@
|
||||
---
|
||||
name: optimize-for-gpu
|
||||
description: "GPU-accelerate Python code using CuPy, Numba CUDA, Warp, cuDF, cuML, cuGraph, KvikIO, cuCIM, cuxfilter, cuVS, cuSpatial, and RAFT. Use whenever the user mentions GPU/CUDA/NVIDIA acceleration, or wants to speed up NumPy, pandas, scikit-learn, scikit-image, NetworkX, GeoPandas, or Faiss workloads. Covers physics simulation, differentiable rendering, mesh ray casting, particle systems (DEM/SPH/fluids), vector/similarity search, GPUDirect Storage file IO, interactive dashboards, geospatial analysis, medical imaging, and sparse eigensolvers. Also use when you see CPU-bound Python code (loops, large arrays, ML pipelines, graph analytics, image processing) that would benefit from GPU acceleration, even if not explicitly requested."
|
||||
description: GPU-accelerates scientific Python on NVIDIA hardware and verifies that the result is correct and faster. Use for CUDA/GPU optimization; CPU-bound NumPy, SciPy, pandas, scikit-learn, NetworkX, scikit-image, vector-search, image-processing, graph, simulation, or file-I/O workloads; CuPy, cuDF, cuML, cuGraph, cuVS, cuCIM, KvikIO, Warp, Newton, Numba-CUDA, or RAFT questions; and profiling, memory-transfer, kernel, or multi-GPU bottlenecks. Also use when large data-parallel Python code is slow and GPU acceleration is a plausible option, even if the user does not name CUDA.
|
||||
license: MIT
|
||||
compatibility: Requires an NVIDIA CUDA-capable GPU for GPU execution. RAPIDS 26.06 requires Python 3.11+ on Linux or WSL2 and matching CUDA 12 or 13 wheels. Package installation needs network access.
|
||||
metadata:
|
||||
version: "1.2"
|
||||
author: K-Dense, Inc.
|
||||
version: "1.3"
|
||||
skill-author: K-Dense, Inc.
|
||||
---
|
||||
|
||||
# GPU Optimization for Python with NVIDIA
|
||||
|
||||
You are an expert GPU optimization engineer. Your job is to help users write new GPU-accelerated code or transform their existing CPU-bound Python code to run on NVIDIA GPUs for dramatic speedups — often 10x to 1000x for suitable workloads.
|
||||
Treat GPU acceleration as an evidence-driven optimization, not an automatic rewrite. Preserve the
|
||||
user's numerical and algorithmic contract, measure with representative data, and keep the GPU
|
||||
version only when synchronized end-to-end benchmarks show a useful improvement.
|
||||
|
||||
## When This Skill Applies
|
||||
|
||||
@@ -36,24 +40,34 @@ You are an expert GPU optimization engineer. Your job is to help users write new
|
||||
- User is doing simulations, signal processing, financial modeling, bioinformatics, physics, or any compute-intensive work
|
||||
- User wants to optimize existing code and GPU acceleration is the right answer
|
||||
|
||||
## Choosing a Library
|
||||
## Choose the Smallest Suitable Layer
|
||||
|
||||
Pick the GPU library by the CPU library it replaces:
|
||||
Prefer a maintained library implementation over a custom kernel:
|
||||
|
||||
| CPU library | GPU replacement | Use for |
|
||||
| Existing workload | Preferred path | Use for |
|
||||
| --- | --- | --- |
|
||||
| NumPy | **CuPy** | array and matrix operations |
|
||||
| (custom loops) | **Numba CUDA** | hand-written GPU kernels |
|
||||
| (simulation) | **Warp** | simulation, spatial computing, differentiable programming |
|
||||
| pandas | **cuDF** | dataframe operations |
|
||||
| scikit-learn | **cuML** | machine learning |
|
||||
| NetworkX | **cuGraph** | graph analytics |
|
||||
| (file IO) | **KvikIO** | high-performance GPU file IO |
|
||||
| (dashboards) | **cuxfilter** | GPU-accelerated interactive dashboards |
|
||||
| scikit-image | **cuCIM** | image processing |
|
||||
| Faiss / Annoy | **cuVS** | vector search |
|
||||
| GeoPandas | **cuSpatial** | geospatial analytics |
|
||||
| (low-level) | **RAFT** (pylibraft) | GPU primitives and multi-GPU |
|
||||
| NumPy / SciPy | **CuPy** | arrays, sparse matrices, linear algebra, FFTs, signal processing |
|
||||
| pandas | **cudf.pandas**, then **cuDF** | accelerator mode first; native API for more control |
|
||||
| scikit-learn | **cuml.accel**, then **cuML** | accelerator mode first; native estimators as needed |
|
||||
| NetworkX | **nx-cugraph**, then **cuGraph** | backend dispatch first; native graph API at scale |
|
||||
| scikit-image | **cuCIM** | GPU image processing and whole-slide imaging |
|
||||
| Faiss / Annoy / k-NN | **cuVS** | exact and approximate vector search |
|
||||
| Raw or remote file I/O | **KvikIO** | GPU buffers and GPUDirect Storage |
|
||||
| Custom array kernels | **Numba-CUDA-MLIR** for new work; **Numba-CUDA** for existing code | explicit SIMT kernels and shared memory |
|
||||
| Spatial or differentiable kernels | **Warp** | geometry, simulation kernels, robotics, autodiff |
|
||||
| High-level physics simulation | **Newton** | maintained engine that succeeds the removed `warp.sim` module |
|
||||
| Low-level RAPIDS primitives | **RAFT** (`pylibraft`) | sparse eigensolvers, resources, multi-GPU building blocks |
|
||||
|
||||
Do not move code out of PyTorch, JAX, TensorFlow, or another GPU-native framework merely to use
|
||||
one of these libraries. First remove CPU round trips and use the framework's compiler, profiler,
|
||||
mixed-precision, and batching facilities.
|
||||
|
||||
Treat these as legacy-only:
|
||||
|
||||
| Project | Status | Guidance |
|
||||
| --- | --- | --- |
|
||||
| **cuxfilter** | Final release 26.06 | Maintain existing dashboards only. For new work, combine cuDF with HoloViews/hvPlot/Datashader and serve with Panel, Dash, Streamlit, or Bokeh. |
|
||||
| **cuSpatial** | Archived at 25.04 | Use only in an isolated legacy environment. For new work, keep geometry in GeoPandas/Shapely and accelerate compatible tabular stages with cuDF. |
|
||||
|
||||
Full per-library guidance, including when each is the *wrong* choice and how to combine
|
||||
them, is in [references/decision_framework.md](references/decision_framework.md).
|
||||
@@ -64,56 +78,81 @@ every library are in
|
||||
|
||||
## Optimization Workflow
|
||||
|
||||
When helping a user optimize code, follow this process:
|
||||
### 1. Define the contract and baseline
|
||||
|
||||
- Capture a representative input, expected output, and acceptable numerical tolerance.
|
||||
- Measure the current end-to-end path, including input, transfers, compute, and output.
|
||||
- Profile before changing code. Use CPU profilers for CPU code and identify whether the real limit
|
||||
is compute, memory bandwidth, allocation, transfer, synchronization, or storage.
|
||||
- Record hardware, package versions, dtypes, shapes, batch size, and warm-up policy with results.
|
||||
|
||||
### 2. Check suitability before porting
|
||||
|
||||
GPU execution is promising when the hot path exposes substantial independent work, runs often
|
||||
enough to amortize initialization and transfer, and has a working set that fits available device
|
||||
memory with room for temporaries. Keep a CPU path when the workload is small, mostly sequential,
|
||||
dominated by unsupported operations, or requires frequent host-device round trips.
|
||||
|
||||
Do not use fixed row-count thresholds as proof. Benchmark the user's actual shapes and hardware.
|
||||
For out-of-core data, estimate peak working memory and choose chunking, Dask, or a streaming design
|
||||
before allocating.
|
||||
|
||||
### 3. Try the least disruptive implementation
|
||||
|
||||
1. If the code already uses a GPU-native framework, optimize within that framework.
|
||||
2. Try accelerator or backend modes (`cudf.pandas`, `cuml.accel`, `nx-cugraph`).
|
||||
3. Move to a native GPU API only where accelerator coverage or performance is insufficient.
|
||||
4. Write a custom kernel only when profiling shows an operation without a suitable library
|
||||
implementation.
|
||||
|
||||
Read the relevant library reference before writing code; compatible names can still differ in
|
||||
defaults, dtypes, output types, and supported arguments.
|
||||
|
||||
### 4. Keep a coherent GPU data path
|
||||
|
||||
- Transfer inputs once and keep intermediates device-resident.
|
||||
- Reuse allocations and prefer `out=` or in-place forms when semantics allow.
|
||||
- Batch small operations; fuse elementwise work when it removes intermediate arrays.
|
||||
- Use pinned host memory and non-default streams only after profiling shows transfer overlap matters.
|
||||
- Choose `float32`, mixed precision, or reduced-precision storage only when the contract permits it.
|
||||
|
||||
### 5. Validate semantics before speed
|
||||
|
||||
- Compare CPU and GPU outputs on small deterministic fixtures and representative data.
|
||||
- Use explicit tolerances for floating-point results and test edge cases, NaNs, ordering, and dtypes.
|
||||
- For approximate nearest-neighbor indexes, report recall@k against exact search; do not compare an
|
||||
exact CPU algorithm with an approximate GPU algorithm as if they were equivalent.
|
||||
- Check accelerator warnings and logs for CPU fallback.
|
||||
|
||||
### 6. Benchmark GPU code correctly
|
||||
|
||||
GPU work is asynchronous, so a CPU timer around an unsynchronized call measures enqueue time.
|
||||
Warm up context creation and JIT compilation, then use CUDA events or a library-aware timer:
|
||||
|
||||
### 1. Profile First
|
||||
Before optimizing, understand where time is actually spent:
|
||||
```python
|
||||
import time
|
||||
# or use cProfile, line_profiler, or py-spy for detailed profiling
|
||||
from cupyx.profiler import benchmark
|
||||
|
||||
print(benchmark(gpu_function, (arg1, arg2), n_warmup=10, n_repeat=100))
|
||||
```
|
||||
Don't guess — measure. The bottleneck might not be where the user thinks.
|
||||
|
||||
### 2. Assess GPU Suitability
|
||||
Not all code benefits from GPU acceleration. GPU excels when:
|
||||
- **Data parallelism is high**: The same operation applies to thousands/millions of elements
|
||||
- **Compute intensity is high**: Many FLOPs per byte of memory accessed
|
||||
- **Data is large enough**: GPU overhead means small arrays (< ~10K elements) may be slower on GPU
|
||||
- **Memory fits**: Data must fit in GPU memory (typically 8-80 GB)
|
||||
Use `%gpu_timeit` in notebooks, Nsight Systems (`nsys`) for end-to-end timelines, and Nsight
|
||||
Compute (`ncu`) for kernel analysis. Report both synchronized kernel/region time and realistic
|
||||
end-to-end latency; include transfer and conversion costs when production pays them.
|
||||
|
||||
GPU is a poor fit when:
|
||||
- Data is tiny (< 10K elements)
|
||||
- Algorithm is inherently sequential with data dependencies between steps
|
||||
- Code is I/O bound (disk, network), not compute bound — though KvikIO with GPUDirect Storage can help when IO feeds GPU compute
|
||||
- Many small, heterogeneous operations (kernel launch overhead dominates)
|
||||
### 7. Keep, revise, or reject the port
|
||||
|
||||
### 3. Start Simple, Then Optimize
|
||||
1. **Try the drop-in replacement first.** CuPy for NumPy, cudf.pandas for pandas, cuml.accel for sklearn, nx-cugraph for NetworkX. This alone often gives 5-50x speedup.
|
||||
2. **Minimize host-device transfers.** Keep data on GPU. Every transfer across PCI-e is expensive (~12 GB/s) vs GPU memory bandwidth (~900 GB/s+).
|
||||
3. **Batch operations.** Fewer large GPU operations beat many small ones.
|
||||
4. **Only write custom kernels if needed.** CuPy and cuDF use NVIDIA's hand-tuned libraries. Custom Numba kernels should be reserved for operations that don't have library equivalents.
|
||||
5. **Profile the GPU version.** Use `nvprof`, `nsys`, or CuPy's built-in benchmarking.
|
||||
|
||||
### 4. Memory Management Principles
|
||||
These apply across all libraries:
|
||||
- **Pre-allocate output arrays** instead of creating new ones in loops
|
||||
- **Reuse GPU memory** — use memory pools (CuPy has this built-in)
|
||||
- **Use pinned (page-locked) host memory** for faster CPU-GPU transfers
|
||||
- **Avoid unnecessary copies** — use in-place operations where possible
|
||||
- **Stream operations** for overlapping compute and data transfer
|
||||
|
||||
### 5. Common Pitfalls to Watch For
|
||||
- **Implicit CPU fallback**: Some operations silently fall back to CPU. Watch for warnings.
|
||||
- **Synchronization overhead**: GPU operations are asynchronous. Calling `.get()` or `cp.asnumpy()` forces a sync.
|
||||
- **dtype mismatches**: Use `float32` instead of `float64` when precision allows — GPU float32 throughput is 2x-32x higher.
|
||||
- **Small kernel launches**: Each kernel launch has ~5-20us overhead. Fuse operations when possible.
|
||||
Retain the GPU path only when it passes correctness checks and improves the metric the user cares
|
||||
about on representative data. If it does not, explain whether the limiting factor is problem size,
|
||||
transfers, unsupported fallback, memory pressure, launch granularity, or the algorithm itself.
|
||||
|
||||
## Important Notes
|
||||
|
||||
- Always handle the case where no GPU is available — provide a CPU fallback or clear error message
|
||||
- Provide a CPU fallback when the application requires portability; otherwise fail early with a
|
||||
clear hardware and dependency error.
|
||||
- Test numerical correctness against CPU results (GPU floating point may differ slightly due to operation ordering)
|
||||
- GPU memory is limited — for datasets larger than GPU memory, consider chunking or using RAPIDS Dask for multi-GPU
|
||||
- The CUDA Array Interface enables zero-copy sharing between CuPy, Numba, Warp, cuDF, cuML, cuGraph, cuVS, cuSpatial, KvikIO, PyTorch, and JAX arrays on GPU
|
||||
- Prefer the CUDA Array Interface or DLPack for supported zero-copy interchange, but verify device,
|
||||
dtype, contiguity, ownership, and stream semantics rather than assuming every conversion is free.
|
||||
|
||||
## Reference Files
|
||||
|
||||
@@ -122,16 +161,16 @@ Before writing any GPU optimization code, read the relevant reference file(s):
|
||||
| File | When to Read |
|
||||
|------|-------------|
|
||||
| `references/cupy.md` | User has NumPy/SciPy code, or needs array operations on GPU |
|
||||
| `references/numba.md` | User needs custom CUDA kernels, fine-grained GPU control, or GPU ufuncs |
|
||||
| `references/numba.md` | User has existing Numba-CUDA code or needs explicit SIMT kernels; note the migration path to Numba-CUDA-MLIR |
|
||||
| `references/cudf.md` | User has pandas code, or needs dataframe operations on GPU |
|
||||
| `references/cuml.md` | User has scikit-learn code, or needs ML training/inference/preprocessing on GPU |
|
||||
| `references/cugraph.md` | User has NetworkX code, or needs graph analytics on GPU |
|
||||
| `references/warp.md` | User needs GPU simulation, spatial computing, mesh/volume queries, differentiable programming, or robotics |
|
||||
| `references/warp.md` | User needs GPU kernels for simulation, spatial computing, mesh/volume queries, differentiable programming, or robotics; use Newton for a high-level physics engine |
|
||||
| `references/kvikio.md` | User needs high-performance file IO to/from GPU, GPUDirect Storage, reading S3/HTTP to GPU, or Zarr on GPU |
|
||||
| `references/cuxfilter.md` | User wants GPU-accelerated interactive dashboards, cross-filtering, or EDA visualization (note: sunset — 26.06 is the final release) |
|
||||
| `references/cuxfilter.md` | User maintains or explicitly requests cuxfilter (sunset — 26.06 is the final release) |
|
||||
| `references/cucim.md` | User has scikit-image code, or needs image processing, digital pathology, or WSI reading on GPU |
|
||||
| `references/cuvs.md` | User needs vector search, nearest neighbors, similarity search, or RAG retrieval on GPU |
|
||||
| `references/cuspatial.md` | User has GeoPandas/shapely code, or needs spatial joins, distance calculations, or trajectory analysis on GPU (note: archived — frozen at 25.04) |
|
||||
| `references/cuspatial.md` | User maintains or explicitly requests cuSpatial (archived — frozen at 25.04 and isolated from current RAPIDS) |
|
||||
| `references/raft.md` | User needs sparse eigensolvers, device memory management, or multi-GPU primitives |
|
||||
|
||||
Read the specific reference before writing code — they contain detailed API patterns, optimization techniques, and pitfalls specific to each library.
|
||||
|
||||
@@ -2,10 +2,9 @@
|
||||
|
||||
Before/after conversions: NumPy to CuPy, pandas to cuDF, a custom loop to a Numba CUDA
|
||||
kernel, NetworkX to cuGraph, scikit-learn to cuML, a simulation loop to a Warp kernel,
|
||||
file IO to KvikIO, dashboards to cuxfilter, scikit-image to cuCIM, GeoPandas to
|
||||
cuSpatial, Faiss/Annoy to cuVS, and `scipy.sparse.linalg` to RAFT.
|
||||
|
||||
## Code Transformation Patterns
|
||||
file IO to KvikIO, maintained GPU-backed dashboards, scikit-image to cuCIM, legacy
|
||||
GeoPandas-to-cuSpatial point-in-polygon, exact Faiss to exact cuVS search, and
|
||||
`scipy.sparse.linalg` to RAFT.
|
||||
|
||||
When converting existing CPU code, apply these patterns:
|
||||
|
||||
@@ -56,9 +55,12 @@ def process(data, out):
|
||||
if i < data.size:
|
||||
out[i] = math.sin(data[i]) * math.exp(-data[i])
|
||||
|
||||
d_data = cuda.to_device(data)
|
||||
d_out = cuda.device_array(d_data.shape, dtype=d_data.dtype)
|
||||
threads = 256
|
||||
blocks = (len(data) + threads - 1) // threads
|
||||
process[blocks, threads](d_data, d_out)
|
||||
out = d_out.copy_to_host()
|
||||
```
|
||||
|
||||
### NetworkX to cuGraph
|
||||
@@ -157,7 +159,12 @@ with kvikio.RemoteFile.open_s3_url("s3://bucket/data.bin") as f:
|
||||
f.read(buf)
|
||||
```
|
||||
|
||||
### GPU-accelerated dashboard with cuxfilter
|
||||
### GPU-backed dashboard with maintained libraries
|
||||
|
||||
cuxfilter ended with RAPIDS 26.06. Do not start a new application with it. Keep large
|
||||
transformations and aggregations in cuDF, then transfer only the compact display data at an
|
||||
explicit visualization boundary:
|
||||
|
||||
```python
|
||||
# Before — static matplotlib/seaborn plots, no interactivity
|
||||
import pandas as pd
|
||||
@@ -169,27 +176,28 @@ df.plot.scatter(x="feature1", y="feature2", ax=axes[0])
|
||||
df["category"].value_counts().plot.bar(ax=axes[1])
|
||||
plt.show()
|
||||
|
||||
# After (GPU) — interactive cross-filtering dashboard
|
||||
# After — GPU data preparation plus a maintained dashboard stack
|
||||
import cudf
|
||||
import cuxfilter
|
||||
import hvplot.pandas # Registers .hvplot on pandas objects
|
||||
import panel as pn
|
||||
|
||||
df = cudf.read_parquet("large_dataset.parquet")
|
||||
cux_df = cuxfilter.DataFrame.from_dataframe(df)
|
||||
|
||||
scatter = cuxfilter.charts.scatter(x="feature1", y="feature2", pixel_shade_type="linear")
|
||||
bar = cuxfilter.charts.bar("category")
|
||||
slider = cuxfilter.charts.range_slider("value_col")
|
||||
|
||||
d = cux_df.dashboard(
|
||||
[scatter, bar],
|
||||
sidebar=[slider],
|
||||
layout=cuxfilter.layouts.feature_and_base,
|
||||
theme=cuxfilter.themes.rapids_dark,
|
||||
title="Interactive Explorer",
|
||||
gpu_df = cudf.read_parquet("large_dataset.parquet")
|
||||
gpu_summary = (
|
||||
gpu_df.groupby("category", as_index=False)
|
||||
.agg({"value_col": "mean"})
|
||||
)
|
||||
d.app() # or d.show() for standalone web app
|
||||
display_summary = gpu_summary.to_pandas() # Transfer only the reduced result
|
||||
dashboard = pn.Column(
|
||||
"# Interactive Explorer",
|
||||
display_summary.hvplot.bar(x="category", y="value_col"),
|
||||
)
|
||||
dashboard.servable()
|
||||
```
|
||||
|
||||
For linked selections over detailed points, use HoloViews/hvPlot with Datashader and Panel.
|
||||
Keep filter/aggregation callbacks on the GPU where practical, and document every conversion to
|
||||
pandas. Read the cuxfilter reference only when maintaining an existing 26.06 application.
|
||||
|
||||
### scikit-image to cuCIM
|
||||
```python
|
||||
# Before (CPU)
|
||||
@@ -218,46 +226,58 @@ labels = label(cleaned)
|
||||
props = regionprops_table(labels, image_gpu, properties=['area', 'centroid'])
|
||||
```
|
||||
|
||||
### GeoPandas to cuSpatial
|
||||
### GeoPandas point-in-polygon to cuSpatial (legacy 25.04 only)
|
||||
|
||||
cuSpatial is archived and incompatible with current RAPIDS packages. Use this only in an isolated
|
||||
environment pinned to 25.04. `point_in_polygon` returns a boolean membership matrix; it is not a
|
||||
drop-in replacement for `geopandas.sjoin`.
|
||||
|
||||
```python
|
||||
# Before (CPU)
|
||||
import geopandas as gpd
|
||||
import numpy as np
|
||||
from shapely.geometry import Point
|
||||
|
||||
points = gpd.GeoDataFrame(geometry=[Point(x, y) for x, y in coords], crs="EPSG:4326")
|
||||
polygons = gpd.read_file("regions.geojson")
|
||||
joined = gpd.sjoin(points, polygons, predicate="within")
|
||||
|
||||
# After (GPU) — convert and use cuSpatial
|
||||
import cuspatial
|
||||
import cudf
|
||||
|
||||
points_cu = cuspatial.from_geopandas(points)
|
||||
polygons_cu = cuspatial.from_geopandas(polygons)
|
||||
joined = cuspatial.point_in_polygon(
|
||||
points_cu.geometry.x, points_cu.geometry.y,
|
||||
polygons_cu.geometry
|
||||
points = gpd.GeoSeries([Point(x, y) for x, y in coords], crs="EPSG:4326")
|
||||
polygons = gpd.read_file("regions.geojson").geometry.iloc[:31]
|
||||
membership_cpu = np.column_stack(
|
||||
[points.within(polygon).to_numpy() for polygon in polygons]
|
||||
)
|
||||
|
||||
# After (GPU, legacy) — same point-by-polygon membership semantics
|
||||
import cuspatial
|
||||
|
||||
points_gpu = cuspatial.from_geopandas(points)
|
||||
polygons_gpu = cuspatial.from_geopandas(polygons)
|
||||
membership_gpu = cuspatial.point_in_polygon(points_gpu, polygons_gpu)
|
||||
```
|
||||
|
||||
### Faiss/Annoy to cuVS
|
||||
### Exact Faiss search to exact cuVS search
|
||||
|
||||
Match algorithmic semantics before benchmarking. Use cuVS brute force for an exact Faiss
|
||||
`IndexFlatL2` baseline; use CAGRA only when approximate results are acceptable and report recall@k
|
||||
against this exact ground truth.
|
||||
|
||||
```python
|
||||
# Before (CPU) — Faiss
|
||||
import faiss
|
||||
import numpy as np
|
||||
|
||||
embeddings = np.random.rand(1_000_000, 128).astype(np.float32)
|
||||
rng = np.random.default_rng(42)
|
||||
embeddings = rng.random((1_000_000, 128), dtype=np.float32)
|
||||
queries = rng.random((1_000, 128), dtype=np.float32)
|
||||
index = faiss.IndexFlatL2(128)
|
||||
index.add(embeddings)
|
||||
distances, neighbors = index.search(queries, k=10)
|
||||
|
||||
# After (GPU) — cuVS CAGRA (orders of magnitude faster)
|
||||
# After (GPU) — cuVS exact brute-force search
|
||||
import cupy as cp
|
||||
from cuvs.neighbors import cagra
|
||||
from cuvs.neighbors import brute_force
|
||||
|
||||
embeddings = cp.random.rand(1_000_000, 128, dtype=cp.float32)
|
||||
index = cagra.build(cagra.IndexParams(), embeddings)
|
||||
distances, neighbors = cagra.search(cagra.SearchParams(), index, queries, k=10)
|
||||
embeddings_gpu = cp.asarray(embeddings)
|
||||
queries_gpu = cp.asarray(queries)
|
||||
index_gpu = brute_force.build(embeddings_gpu, metric="sqeuclidean")
|
||||
distances_gpu, neighbors_gpu = brute_force.search(index_gpu, queries_gpu, k=10)
|
||||
```
|
||||
|
||||
### scipy.sparse.linalg to RAFT
|
||||
|
||||
@@ -1,6 +1,9 @@
|
||||
# cuCIM Reference
|
||||
|
||||
cuCIM (CUDA Clara IMage) is NVIDIA's GPU-accelerated computer vision and image processing library within the RAPIDS ecosystem. Its `cucim.skimage` module is a near-drop-in GPU replacement for scikit-image, with 200+ GPU-accelerated functions. It also provides a high-performance whole-slide image (WSI) reader via `cucim.clara.CuImage` that is 5-6x faster than OpenSlide.
|
||||
cuCIM (CUDA Clara IMage) is NVIDIA's GPU-accelerated computer vision and image processing library
|
||||
within the RAPIDS ecosystem. Its `cucim.skimage` module mirrors a substantial part of scikit-image
|
||||
on CuPy arrays, and `cucim.CuImage` reads tiled whole-slide images. Verify function coverage and
|
||||
benchmark the actual image sizes, storage path, and processing chain.
|
||||
|
||||
> **Full documentation:** https://docs.rapids.ai/api/cucim/stable/
|
||||
> **GitHub:** https://github.com/rapidsai/cucim
|
||||
@@ -33,11 +36,12 @@ cuCIM (CUDA Clara IMage) is NVIDIA's GPU-accelerated computer vision and image p
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cucim-cu12 # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cucim-cu13 # For CUDA 13.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cucim-cu12==26.6.*" # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cucim-cu13==26.6.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
cuCIM wheels are also published directly to PyPI, so the extra index is optional.
|
||||
@@ -475,7 +479,7 @@ distances = distance_transform_edt(binary_image_gpu)
|
||||
|
||||
## Whole-Slide Image Reading
|
||||
|
||||
`cucim.clara.CuImage` — high-performance WSI reader, compatible with OpenSlide API, 5-6x faster.
|
||||
`cucim.CuImage` is the public whole-slide image reader:
|
||||
|
||||
```python
|
||||
from cucim import CuImage
|
||||
@@ -513,34 +517,27 @@ cache = ImageCache(memory_capacity=2 * 1024**3) # 2 GB cache
|
||||
|
||||
### GPUDirect Storage
|
||||
|
||||
For large files (2GB+), GPUDirect Storage bypasses CPU memory for 25%+ additional speedup:
|
||||
|
||||
```python
|
||||
from cucim.clara.filesystem import CuFileDriver
|
||||
|
||||
# Read directly into GPU memory, bypassing CPU
|
||||
driver = CuFileDriver(path, flags)
|
||||
driver.pread(gpu_buffer, size, offset)
|
||||
```
|
||||
GPUDirect Storage can reduce CPU staging on a supported Linux, driver, filesystem, storage, and
|
||||
container configuration. Treat it as a deployment capability, not an automatic size-based
|
||||
optimization. Confirm that GDS is active and compare end-to-end tile throughput; otherwise use
|
||||
cuCIM's normal reader path. Use KvikIO for explicit raw-buffer I/O rather than depending on
|
||||
cuCIM-internal filesystem classes.
|
||||
|
||||
---
|
||||
|
||||
## Performance Characteristics
|
||||
|
||||
**Headline numbers:**
|
||||
- Up to **1245x faster** than scikit-image for certain operations on large images
|
||||
- **5-6x faster** than OpenSlide for WSI multi-threaded patch reading
|
||||
- **25%+ additional speedup** with GPUDirect Storage on 2GB+ files
|
||||
Benchmark each operation and the complete pipeline. Image dimensions alone do not predict a
|
||||
speedup: kernel type, footprint size, channels, dtype, storage, tiling, transfer, and downstream
|
||||
reuse all matter.
|
||||
|
||||
**Scaling behavior:**
|
||||
- **4K resolution and above:** GPU parallelism fully utilized, maximum speedups
|
||||
- **~1000x1000:** Moderate but measurable speedups for most operations
|
||||
- **Below ~512x512:** Diminishing returns; GPU overhead starts to matter
|
||||
- **Below ~64x64:** CPU may be faster due to CUDA kernel launch overhead
|
||||
|
||||
**First-call overhead:** JIT compilation on first kernel execution (cached after). Benchmark on subsequent calls.
|
||||
|
||||
**Best strategy:** Transfer image to GPU once, chain all processing operations, transfer back once at the end.
|
||||
1. Warm the CUDA context and lazy kernels before timed repetitions.
|
||||
2. Compare equivalent border modes, interpolation, connectivity, dtype, and output semantics.
|
||||
3. Report decode/read, host-device transfer, processing, and end-to-end times separately.
|
||||
4. For WSI workloads, record tile size, level, compression, access pattern, worker count, cache
|
||||
state, storage device, and whether GDS was active.
|
||||
5. Transfer an image to the GPU once, chain compatible operations, and transfer only the result
|
||||
required by the next CPU consumer.
|
||||
|
||||
---
|
||||
|
||||
@@ -577,7 +574,8 @@ result = gaussian(cp.asarray(image), sigma=5)
|
||||
|
||||
4. **Data must be explicitly moved to GPU.** cuCIM does not auto-transfer; you must call `cp.asarray()`.
|
||||
|
||||
5. **Small image penalty.** Images below ~512x512 may not benefit. Below ~64x64, CPU is likely faster.
|
||||
5. **Small image penalty.** Small or one-shot operations may not amortize context, launch, and
|
||||
transfer overhead. Benchmark the real tile size and batching strategy.
|
||||
|
||||
6. **GPU memory constraints.** Very large images must be tiled. GPU memory is typically smaller than system RAM.
|
||||
|
||||
|
||||
@@ -27,11 +27,12 @@ cuDF is a GPU DataFrame library that provides a pandas-like API for loading, joi
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add cudf-cu12 # For CUDA 12.x
|
||||
uv add cudf-cu13 # For CUDA 13.x
|
||||
uv add "cudf-cu12==26.6.*" # For CUDA 12.x
|
||||
uv add "cudf-cu13==26.6.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
cuDF wheels are now published directly to PyPI — the `--extra-index-url=https://pypi.nvidia.com` extra index is no longer required. Requires Python >= 3.11.
|
||||
|
||||
@@ -1,6 +1,9 @@
|
||||
# cuGraph Reference
|
||||
|
||||
cuGraph is NVIDIA's GPU-accelerated graph analytics library within the RAPIDS ecosystem. It provides NetworkX-compatible APIs for graph algorithms, delivering 10-500x+ speedup over CPU-based NetworkX on medium to large graphs. It supports both a direct Python API and a **zero-code-change NetworkX backend** (nx-cugraph) that accelerates existing NetworkX code with no modifications.
|
||||
cuGraph is NVIDIA's GPU-accelerated graph analytics library within the RAPIDS ecosystem. It
|
||||
supports both a direct Python API and an **nx-cugraph** NetworkX backend. Performance depends on
|
||||
algorithm, graph topology and size, graph-construction cost, fallback behavior, and GPU hardware,
|
||||
so benchmark the complete workload instead of promising a fixed speedup.
|
||||
|
||||
> **Full documentation:** https://docs.rapids.ai/api/cugraph/stable/
|
||||
> **Version (stable):** 26.06.00
|
||||
@@ -27,11 +30,12 @@ cuGraph is NVIDIA's GPU-accelerated graph analytics library within the RAPIDS ec
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cugraph-cu12 # Core cuGraph for CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com nx-cugraph-cu12 # NetworkX backend
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cugraph-cu12==26.6.*" # Core cuGraph for CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "nx-cugraph-cu12==26.6.*" # NetworkX backend
|
||||
# For CUDA 13.x, use the -cu13 packages: cugraph-cu13, nx-cugraph-cu13
|
||||
```
|
||||
|
||||
@@ -485,7 +489,7 @@ cugraph-pyg provides native GPU-accelerated implementations of PyG's core interf
|
||||
- **Sampler/Loader**: GPU-accelerated neighborhood sampling with configurable fan-out
|
||||
|
||||
```bash
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cugraph-pyg-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cugraph-pyg-cu12==26.6.*"
|
||||
```
|
||||
|
||||
**Key capabilities:**
|
||||
@@ -501,7 +505,7 @@ uv add --extra-index-url=https://pypi.nvidia.com cugraph-pyg-cu12
|
||||
WholeGraph provides distributed GPU memory management for large-scale GNN training through its **WholeMemory** abstraction.
|
||||
|
||||
```bash
|
||||
uv add --extra-index-url=https://pypi.nvidia.com pylibwholegraph-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "pylibwholegraph-cu12==26.6.*"
|
||||
```
|
||||
|
||||
**Core concepts:**
|
||||
@@ -534,45 +538,23 @@ uv add --extra-index-url=https://pypi.nvidia.com pylibwholegraph-cu12
|
||||
|
||||
## Performance Characteristics and Benchmarks
|
||||
|
||||
### nx-cugraph Benchmarks (NetworkX backend)
|
||||
Benchmark with the user's actual topology and algorithm:
|
||||
|
||||
**Hardware:** Intel Xeon w9-3495X (56 cores), NVIDIA RTX 3090 (24GB), 251 GB RAM, CUDA 12.8
|
||||
1. Check that the requested NetworkX algorithm dispatches to nx-cugraph rather than falling back
|
||||
to CPU.
|
||||
2. Measure graph construction and conversion separately from repeated algorithm calls, then report
|
||||
both warm algorithm time and end-to-end time.
|
||||
3. Warm the CUDA context before collecting timed repetitions.
|
||||
4. Verify output semantics, convergence tolerance, sampled-algorithm parameters, and floating-point
|
||||
tolerance against the CPU implementation.
|
||||
5. Record vertices, edges, directedness, weight dtype, GPU model, CPU baseline, software versions,
|
||||
and whether the graph was already device-resident.
|
||||
6. Use multi-GPU only after estimating graph and temporary-memory requirements and measuring the
|
||||
communication cost.
|
||||
|
||||
**Datasets tested:**
|
||||
|
||||
| Dataset | Nodes | Edges | Type |
|
||||
|---|---|---|---|
|
||||
| netscience | 1,461 | 5,484 | Small |
|
||||
| amazon0302 | 262,111 | 1,234,877 | Medium |
|
||||
| cit-Patents | 3,774,768 | 16,518,948 | Large |
|
||||
| soc-LiveJournal1 | 4,847,571 | 68,993,773 | Very large |
|
||||
|
||||
**Speedups (GPU vs CPU NetworkX):**
|
||||
|
||||
| Algorithm | Medium Graph | Large Graph | Very Large Graph |
|
||||
|---|---|---|---|
|
||||
| `betweenness_centrality` (k=100) | ~20x | ~520x | ~300x |
|
||||
| `katz_centrality` | ~100x | ~5,000x | ~24,768x |
|
||||
| `average_clustering` | ~50x | ~1,000x | ~2,828x |
|
||||
| `transitivity` | ~50x | ~1,000x | ~2,832x |
|
||||
| `louvain_communities` | ~30x | ~273x | ~200x |
|
||||
| `pagerank` | ~2x | ~50x | ~188x |
|
||||
| `eigenvector_centrality` | ~7x | ~100x | ~376x |
|
||||
| `k_truss` | ~8x | ~200x | ~540x |
|
||||
|
||||
**Key finding:** Speedup increases dramatically with graph size. Small graphs (< 5K edges) may see overhead from GPU initialization that negates speedup. For graphs with > 100K edges, expect 10-500x+ improvement on most algorithms.
|
||||
|
||||
**Concrete example:** Betweenness centrality on cit-Patents (3.7M nodes, 16.5M edges):
|
||||
- CPU NetworkX: 7 min 41 sec
|
||||
- nx-cugraph GPU: 5.32 sec (~86x speedup)
|
||||
|
||||
### General Performance Guidelines
|
||||
|
||||
- **Small graphs (< 10K edges):** GPU overhead may dominate; NetworkX CPU may be faster
|
||||
- **Medium graphs (100K-1M edges):** 10-100x speedup typical
|
||||
- **Large graphs (1M-100M edges):** 100-1000x+ speedup typical
|
||||
- **Very large graphs (> 100M edges):** Use multi-GPU; single GPU memory may be insufficient
|
||||
- **First call overhead:** Initial GPU kernel compilation and graph transfer adds ~1-3 seconds; subsequent calls on same graph are much faster
|
||||
Small or one-shot graphs can lose to NetworkX after setup and transfer. Large device-resident
|
||||
graphs with supported algorithms are better candidates, but graph size alone does not establish a
|
||||
speedup.
|
||||
|
||||
---
|
||||
|
||||
@@ -634,13 +616,12 @@ import cupy, scipy
|
||||
### With NetworkX
|
||||
```python
|
||||
import networkx as nx
|
||||
import cugraph
|
||||
import nx_cugraph as nxcg
|
||||
|
||||
# NetworkX -> cuGraph
|
||||
# Convert once when repeated algorithms justify keeping the graph on GPU.
|
||||
G_nx = nx.karate_club_graph()
|
||||
G_cu = cugraph.from_networkx(G_nx) # Not yet available in all versions
|
||||
|
||||
# Or use nx-cugraph backend for transparent acceleration
|
||||
G_gpu = nxcg.from_networkx(G_nx)
|
||||
result = nx.pagerank(G_gpu)
|
||||
```
|
||||
|
||||
### With PyTorch Geometric
|
||||
|
||||
@@ -1,6 +1,9 @@
|
||||
# cuML Reference
|
||||
|
||||
cuML is NVIDIA's GPU-accelerated machine learning library within the RAPIDS ecosystem. It provides scikit-learn-compatible APIs for 50+ algorithms, delivering 10-50x faster performance on average, with some algorithms (HDBSCAN, t-SNE, UMAP, KNN) reaching 60-600x speedup. It follows the familiar fit/predict/transform pattern from sklearn.
|
||||
cuML is NVIDIA's GPU-accelerated machine learning library within the RAPIDS ecosystem. It
|
||||
provides scikit-learn-compatible APIs for classification, regression, clustering, dimensionality
|
||||
reduction, preprocessing, and model selection. Performance depends on algorithm, shape, dtype,
|
||||
fallback behavior, and transfer cost; benchmark the complete pipeline on representative data.
|
||||
|
||||
> **Full documentation:** https://docs.rapids.ai/api/cuml/stable/
|
||||
|
||||
@@ -28,11 +31,12 @@ cuML is NVIDIA's GPU-accelerated machine learning library within the RAPIDS ecos
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add cuml-cu12 # For CUDA 12.x
|
||||
uv add cuml-cu13 # For CUDA 13.x
|
||||
uv add "cuml-cu12==26.6.*" # For CUDA 12.x
|
||||
uv add "cuml-cu13==26.6.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
cuML wheels are published directly to PyPI (since RAPIDS 25.10) — the `--extra-index-url=https://pypi.nvidia.com` extra index is no longer required.
|
||||
@@ -407,7 +411,9 @@ X, y = make_regression(n_samples=100_000, n_features=50, noise=0.1)
|
||||
|
||||
## Forest Inference Library
|
||||
|
||||
FIL provides high-performance GPU inference for tree-based models trained in any framework — 80x+ faster than sklearn inference.
|
||||
FIL provides GPU inference for supported tree-based models trained in other frameworks. Its value
|
||||
depends on model structure and inference batch size, so compare warm and end-to-end latency with
|
||||
the deployment baseline.
|
||||
|
||||
```python
|
||||
from cuml.fil import ForestInference
|
||||
@@ -418,7 +424,7 @@ fil_model = ForestInference.load("xgboost_model.ubj", is_classifier=True)
|
||||
# Optional: optimize for specific batch size
|
||||
fil_model.optimize()
|
||||
|
||||
# Predict (80x+ faster than sklearn)
|
||||
# Predict on GPU
|
||||
predictions = fil_model.predict(X_test)
|
||||
probas = fil_model.predict_proba(X_test)
|
||||
```
|
||||
@@ -536,33 +542,39 @@ cuml.accel uses managed memory by default (host RAM augments GPU VRAM). Disable
|
||||
|
||||
## Performance Optimization
|
||||
|
||||
### Expected Speedups by Algorithm
|
||||
### Benchmarking
|
||||
|
||||
| Category | Typical Speedup | Notes |
|
||||
|----------|----------------|-------|
|
||||
| HDBSCAN, t-SNE, UMAP | 60-300x | Complex algorithms benefit most |
|
||||
| KNN | Up to 600x | Scales dramatically with data size |
|
||||
| KMeans, Random Forest | 15-80x | RF: 20-45x single GPU |
|
||||
| FIL inference | 80x+ | Tree model inference from any framework |
|
||||
| Linear models, PCA, Ridge | 2-10x | Simpler algorithms, lower but consistent gains |
|
||||
1. Verify whether `cuml.accel` used a GPU implementation or fell back to scikit-learn.
|
||||
2. Compare the same estimator parameters, train/test split, random seed, and output semantics.
|
||||
3. Warm the CUDA context and any lazy compilation before timed repetitions.
|
||||
4. Report fit, transform/predict, and end-to-end times separately, including conversion and
|
||||
transfer costs paid by the application.
|
||||
5. Record rows, features, sparsity, dtype, estimator parameters, CPU/GPU models, and software
|
||||
versions.
|
||||
6. For stochastic or approximate algorithms, compare quality metrics as well as time.
|
||||
|
||||
### Key Optimization Tips
|
||||
|
||||
1. **Use float32.** GPU float32 throughput is 2x-32x higher than float64. Most ML algorithms don't need double precision.
|
||||
1. **Use float32 when the model's accuracy and stability permit it.** Validate metrics after
|
||||
changing precision; architecture-specific throughput ratios are not a correctness argument.
|
||||
|
||||
2. **Keep data on GPU.** Pass CuPy arrays or cuDF DataFrames. Every NumPy/pandas conversion triggers a device-host transfer.
|
||||
|
||||
3. **Larger datasets = larger speedup.** GPU parallelism advantage grows with data size. Minimum ~10K rows to see benefit.
|
||||
3. **Use enough work to amortize setup and transfer.** Do not rely on a fixed row threshold;
|
||||
features, sparsity, estimator, batch size, and reuse all matter.
|
||||
|
||||
4. **Wide data benefits more.** 128-512 features see higher speedups than 8-16 features.
|
||||
4. **Benchmark the actual feature shape.** Width affects compute, memory traffic, and temporary
|
||||
storage differently for each estimator.
|
||||
|
||||
5. **First call has JIT overhead.** Benchmark on subsequent calls, not the first.
|
||||
|
||||
6. **Use RMM pools.** Pre-allocated memory pools are 1000x faster than raw cudaMalloc.
|
||||
6. **Use RMM pools when allocation overhead or fragmentation is visible in profiling.** Pools
|
||||
amortize allocator costs but reserve device memory and should be sized deliberately.
|
||||
|
||||
7. **Use dask-ml for hyperparameter tuning,** not sklearn's GridSearchCV — it avoids excessive CPU-GPU transfers.
|
||||
|
||||
8. **Use FIL for tree model inference.** Even if the model was trained on CPU (XGBoost, LightGBM, sklearn RF), FIL gives 80x+ inference speedup.
|
||||
8. **Evaluate FIL for supported tree-model inference.** Benchmark the target model and production
|
||||
batch sizes against the existing serving path.
|
||||
|
||||
---
|
||||
|
||||
@@ -692,7 +704,7 @@ print(f"Accuracy: {model.score(X_test, y_test):.4f}")
|
||||
```python
|
||||
from cuml.fil import ForestInference
|
||||
|
||||
# Load XGBoost/LightGBM/sklearn model for 80x+ faster inference
|
||||
# Load a supported XGBoost/LightGBM/sklearn model for GPU inference
|
||||
fil_model = ForestInference.load("my_xgboost_model.ubj", is_classifier=True)
|
||||
predictions = fil_model.predict(X_test)
|
||||
```
|
||||
|
||||
@@ -25,11 +25,12 @@ CuPy is a NumPy/SciPy-compatible array library for GPU-accelerated computing. It
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add cupy-cuda12x # For CUDA 12.x
|
||||
uv add cupy-cuda13x # For CUDA 13.x
|
||||
uv add "cupy-cuda12x==14.1.*" # For CUDA 12.x
|
||||
uv add "cupy-cuda13x==14.1.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
CuPy v14 (current) requires CUDA >= 12.0, Python >= 3.10, and NumPy >= 2.0 (it follows NumPy 2 type-promotion rules, NEP 50), and supports free-threaded Python. The `[ctk]` extra (e.g. `cupy-cuda13x[ctk]`) pulls the required CUDA runtime components from PyPI, so only the NVIDIA driver needs to be pre-installed.
|
||||
@@ -375,8 +376,10 @@ When using CuPy alongside cuDF/RAPIDS, align on a single allocator:
|
||||
|
||||
```python
|
||||
import rmm
|
||||
from rmm.allocators.cupy import rmm_cupy_allocator
|
||||
|
||||
rmm.reinitialize(pool_allocator=True)
|
||||
cp.cuda.set_allocator(rmm.rmm_cupy_allocator)
|
||||
cp.cuda.set_allocator(rmm_cupy_allocator)
|
||||
```
|
||||
|
||||
---
|
||||
@@ -481,7 +484,7 @@ print(result) # Shows CPU and GPU elapsed times with statistics
|
||||
|
||||
In IPython/Jupyter:
|
||||
```python
|
||||
%load_ext cupy
|
||||
%load_ext cupyx.profiler
|
||||
%gpu_timeit my_function(args)
|
||||
```
|
||||
|
||||
@@ -498,19 +501,23 @@ export CUPY_ACCELERATORS=cub # CUB only (default)
|
||||
export CUPY_ACCELERATORS=cub,cutensor # Both (requires cuTENSOR installed)
|
||||
```
|
||||
|
||||
CUB accelerates: reductions (`sum`, `prod`, `amin`, `amax`, `argmin`, `argmax`), inclusive scans (`cumsum`), histograms, sparse matrix-vector multiply, and `ReductionKernel`. Can provide ~100x speedup for reductions.
|
||||
CUB accelerates reductions (`sum`, `prod`, `amin`, `amax`, `argmin`, `argmax`),
|
||||
inclusive scans (`cumsum`), histograms, sparse matrix-vector multiply, and `ReductionKernel`.
|
||||
The benefit depends on dtype, shape, axis, and hardware; benchmark the target operation.
|
||||
|
||||
cuTENSOR accelerates: binary elementwise ufuncs, reduction, tensor contraction.
|
||||
|
||||
### Key Optimization Strategies
|
||||
|
||||
1. **Prefer float32 over float64.** Consumer GPUs have 2x-32x higher float32 throughput. Use `dtype=cp.float32` when precision allows.
|
||||
1. **Prefer float32 over float64 when the numerical contract allows it.** Throughput differences
|
||||
depend on GPU architecture; validate accuracy and benchmark the target device.
|
||||
|
||||
2. **Minimize CPU-GPU transfers.** Every `cp.asnumpy()` / `.get()` triggers synchronization and PCI-e transfer. Keep data on GPU as long as possible.
|
||||
|
||||
3. **Use kernel fusion.** `@cp.fuse()` combines multiple elementwise operations into one kernel, eliminating intermediate arrays.
|
||||
|
||||
4. **Batch operations.** Fewer large operations beat many small ones (kernel launch overhead ~5-20us each).
|
||||
4. **Batch operations.** Fewer large operations usually beat many small ones. Measure launch
|
||||
overhead on the target system rather than relying on a fixed latency.
|
||||
|
||||
5. **Pre-allocate output arrays.** Use `out=` parameter in ufuncs to avoid repeated allocation:
|
||||
```python
|
||||
|
||||
@@ -27,10 +27,11 @@ cuSpatial is a GPU-accelerated GIS library that provides spatial indexing, spati
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cuspatial-cu12 # For CUDA 12.x — installs the final 25.04 release
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuspatial-cu12==25.4.*" # Final 25.04 release
|
||||
```
|
||||
|
||||
The `--extra-index-url=https://pypi.nvidia.com` index is **required** here — the `cuspatial-cu12` entry on PyPI itself is only a stub sdist; the real wheels live on pypi.nvidia.com. There are no CUDA 13 (`-cu13`) packages — the project was archived before CUDA 13 wheels were introduced. Installing cuSpatial pulls in `cudf-cu12==25.4.*` and related 25.04 pins.
|
||||
|
||||
@@ -1,6 +1,9 @@
|
||||
# cuVS Reference
|
||||
|
||||
cuVS is NVIDIA's GPU-accelerated library for vector search and clustering, part of the RAPIDS ecosystem. It provides state-of-the-art implementations of approximate nearest neighbor (ANN) search algorithms on the GPU, delivering orders-of-magnitude speedups over CPU-based libraries like Faiss (CPU mode), Annoy, and scikit-learn's NearestNeighbors for high-dimensional vector search.
|
||||
cuVS is NVIDIA's GPU-accelerated library for exact and approximate nearest-neighbor search, part
|
||||
of the RAPIDS ecosystem. It provides CAGRA, IVF-Flat, IVF-PQ, brute-force, and CPU-serving
|
||||
interoperability. Compare equal metrics and exactness requirements, validate ANN recall, and
|
||||
benchmark build, search, transfer, and serialization costs separately.
|
||||
|
||||
> **Full documentation:** https://docs.rapids.ai/api/cuvs/stable/
|
||||
|
||||
@@ -25,11 +28,12 @@ cuVS is NVIDIA's GPU-accelerated library for vector search and clustering, part
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cuvs-cu12 # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cuvs-cu13 # For CUDA 13.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuvs-cu12==26.6.*" # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuvs-cu13==26.6.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
cuVS wheels (including the companion `libcuvs` wheel) are also published directly to PyPI, so the extra index is optional — but the official cuVS docs still show it and it does no harm.
|
||||
@@ -60,7 +64,8 @@ cuVS is the right tool when the user needs:
|
||||
|
||||
cuVS is NOT the right tool for:
|
||||
- General machine learning (use cuML instead)
|
||||
- Low-dimensional data (< ~16 dimensions) with small datasets (< 10K vectors)
|
||||
- Small or low-dimensional datasets where build, transfer, or launch overhead dominates in a
|
||||
representative benchmark
|
||||
- CPU-only environments with no GPU available
|
||||
|
||||
---
|
||||
@@ -69,13 +74,15 @@ cuVS is NOT the right tool for:
|
||||
|
||||
| Index | Best For | Build Speed | Search Speed | Memory | Accuracy |
|
||||
|-------|----------|-------------|--------------|--------|----------|
|
||||
| **CAGRA** | Default choice — fast build and search | Fast | Fastest | Medium | High |
|
||||
| **IVF-Flat** | When exact distances matter | Medium | Fast | High (stores full vectors) | Very High |
|
||||
| **CAGRA** | Strong first ANN candidate | Fast | Fast | Medium | Tunable |
|
||||
| **IVF-Flat** | High-recall ANN without vector compression | Medium | Fast | High (stores full vectors) | Tunable |
|
||||
| **IVF-PQ** | Large datasets that don't fit in GPU memory | Medium | Fast | Low (compressed) | Good |
|
||||
| **Brute Force** | Small datasets or ground truth | N/A | Slow at scale | High | Exact |
|
||||
| **HNSW** | CPU-side serving from GPU-built index | Slow | Fast (CPU) | Medium | High |
|
||||
|
||||
**Start with CAGRA** unless you have a specific reason not to. It's the fastest GPU-native algorithm and works well for most use cases. Use IVF-PQ when memory is tight, IVF-Flat when you need higher accuracy, and brute force for small datasets or validation.
|
||||
Benchmark CAGRA as the first ANN candidate, then compare IVF-PQ when memory is tight and IVF-Flat
|
||||
when avoiding vector compression matters. Use brute force for exact search and recall ground truth.
|
||||
Choose build and search parameters from measured latency, throughput, memory, and recall targets.
|
||||
|
||||
---
|
||||
|
||||
@@ -393,7 +400,8 @@ Using `float16` halves memory and can speed up both build and search when full f
|
||||
- IVF-Flat/IVF-PQ: increase `n_probes` (default 20)
|
||||
- HNSW: increase `ef` (default 200)
|
||||
|
||||
3. **Use float16 for embeddings.** Most embedding models output float32 but the extra precision rarely matters for similarity search. Cast to float16 to double throughput.
|
||||
3. **Evaluate reduced precision, do not assume it is free.** If the chosen index supports float16,
|
||||
compare recall/ranking metrics and end-to-end performance with float32 before changing storage.
|
||||
|
||||
4. **n_lists tuning for IVF indexes.** A good starting point is `sqrt(n_samples)`. Too few lists = slow search, too many = poor recall.
|
||||
|
||||
|
||||
@@ -29,11 +29,12 @@ cuxfilter is a GPU-accelerated cross-filtering dashboard library from the NVIDIA
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cuxfilter-cu12 # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cuxfilter-cu13 # For CUDA 13.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuxfilter-cu12==26.6.*" # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuxfilter-cu13==26.6.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
Both install the final 26.06 release — no further updates will be published. cuxfilter wheels are also on PyPI directly, so the extra index is optional. cuxfilter depends on cuDF, so `cudf-cu12` (or `cudf-cu13`) will be pulled in automatically.
|
||||
@@ -95,7 +96,7 @@ import cugraph
|
||||
|
||||
edges = cudf.DataFrame({"source": [0, 1, 2], "target": [1, 2, 3], "weight": [1.0, 2.0, 3.0]})
|
||||
G = cugraph.Graph()
|
||||
G.from_cudf_edgelist(edges, destination="target")
|
||||
G.from_cudf_edgelist(edges, source="source", destination="target", edge_attr="weight")
|
||||
cux_df = cuxfilter.DataFrame.load_graph((G.nodes(), G.edges()))
|
||||
```
|
||||
|
||||
@@ -421,7 +422,7 @@ edges = cudf.DataFrame({
|
||||
"target": [1, 2, 2, 3, 3]
|
||||
})
|
||||
G = cugraph.Graph()
|
||||
G.from_cudf_edgelist(edges, destination="target")
|
||||
G.from_cudf_edgelist(edges, source="source", destination="target")
|
||||
|
||||
# Load into cuxfilter (needs node positions — use force_atlas2 or similar layout)
|
||||
positions = cugraph.force_atlas2(G)
|
||||
|
||||
@@ -4,10 +4,13 @@ Each RAPIDS/GPU library, the CPU library it replaces, what it is good at, and wh
|
||||
the wrong choice — CuPy, Numba CUDA, Warp, cuDF, cuML, cuGraph, KvikIO, cuxfilter, cuCIM,
|
||||
cuVS, cuSpatial, and RAFT — plus guidance on combining them.
|
||||
|
||||
## Decision Framework: Which Library to Use
|
||||
|
||||
Choose the right tool based on what the user's code actually does. Read the appropriate reference file(s) before writing any GPU code.
|
||||
|
||||
First decide whether a port is justified: establish an end-to-end baseline, estimate peak device
|
||||
memory including temporaries, and identify transfer and fallback boundaries. Prefer an accelerator
|
||||
or backend mode before a native rewrite, and prefer a maintained library operation before a custom
|
||||
kernel. Validate output semantics and use synchronized GPU timing on representative data.
|
||||
|
||||
### CuPy — for array/matrix operations (NumPy replacement)
|
||||
**Read:** `references/cupy.md`
|
||||
|
||||
@@ -20,18 +23,20 @@ CuPy wraps NVIDIA's optimized libraries (cuBLAS, cuFFT, cuSOLVER, cuSPARSE, cuRA
|
||||
|
||||
**Best for:** Linear algebra, FFTs, array math, image processing, signal processing, Monte Carlo with array ops, any NumPy-heavy workflow.
|
||||
|
||||
### Numba CUDA — for custom GPU kernels
|
||||
### Numba-CUDA-MLIR / Numba-CUDA — for custom GPU kernels
|
||||
**Read:** `references/numba.md`
|
||||
|
||||
Use Numba when the user needs:
|
||||
Use this path when the user needs:
|
||||
- Custom algorithms that don't map to standard array operations
|
||||
- Fine-grained control over GPU threads, blocks, and shared memory
|
||||
- Element-wise operations with complex logic (use `@vectorize(target='cuda')`)
|
||||
- Reduction operations with custom logic
|
||||
- Stencil computations or neighbor-dependent calculations
|
||||
- Anything requiring the CUDA programming model directly
|
||||
|
||||
Numba compiles Python directly into CUDA kernels. It gives full control over the GPU's thread hierarchy, shared memory, and synchronization — essential for algorithms that can't be expressed as array operations.
|
||||
For new projects, evaluate **Numba-CUDA-MLIR**, where NVIDIA is doing new feature development.
|
||||
Use the established `numba-cuda` package for existing `numba.cuda` code, compatibility, or features
|
||||
not yet available in the MLIR implementation; it is in maintenance mode through the CUDA 13
|
||||
lifetime. Do not invest in custom kernels until profiling rules out CuPy or another tuned library.
|
||||
|
||||
**Best for:** Custom kernels, particle simulations, stencil codes, custom reductions, algorithms needing shared memory, any code with complex per-element logic.
|
||||
|
||||
@@ -46,7 +51,11 @@ Use Warp when the user's code is primarily:
|
||||
- Any Python simulation loop that needs to be JIT-compiled to GPU
|
||||
- Spatial computing with meshes, volumes (NanoVDB), hash grids, or BVH queries
|
||||
|
||||
Warp JIT-compiles `@wp.kernel` Python functions to CUDA, with built-in types for spatial computing (vec3, mat33, quat, transform) and primitives for geometry queries (Mesh, Volume, HashGrid, BVH). All kernels are automatically differentiable. Note: the higher-level `warp.sim` module was removed in Warp 1.10 — its functionality moved to the separate Newton physics engine. Warp itself remains the right tool for writing custom simulation kernels.
|
||||
Warp JIT-compiles `@wp.kernel` Python functions to CUDA, with built-in types for spatial computing
|
||||
(vec3, mat33, quat, transform) and primitives for geometry queries (Mesh, Volume, HashGrid, BVH).
|
||||
Warp can generate adjoint kernels for differentiable programs. The higher-level `warp.sim` module
|
||||
was removed in Warp 1.10; use the separate **Newton** engine for maintained high-level rigid-body,
|
||||
robotics, and simulation-environment APIs, and use Warp for custom kernels and domain primitives.
|
||||
|
||||
**Best for:** Physics simulation, mesh ray casting, particle systems, differentiable rendering, robotics kinematics, SDF operations, any workload combining spatial data structures with GPU compute.
|
||||
|
||||
@@ -75,7 +84,10 @@ Use cuML when the user's code is primarily:
|
||||
- Tree model inference (XGBoost, LightGBM, sklearn Random Forest via FIL)
|
||||
- UMAP, t-SNE, HDBSCAN, or KNN on large datasets
|
||||
|
||||
cuML's `cuml.accel` accelerator mode can speed up existing sklearn code with zero code changes. For maximum performance, use the native cuML API. Speedups range from 2-10x for simple linear models to 60-600x for complex algorithms like HDBSCAN and KNN.
|
||||
Start with cuML's `cuml.accel` accelerator mode when compatibility permits. Move to the native
|
||||
cuML API for unsupported estimators, explicit output control, or a measured performance reason.
|
||||
Benchmark the user's estimator, dimensions, and end-to-end pipeline rather than relying on headline
|
||||
speedup ranges.
|
||||
|
||||
**Best for:** Classification, regression, clustering, dimensionality reduction, preprocessing pipelines, model inference, any scikit-learn-heavy workflow.
|
||||
|
||||
@@ -88,7 +100,9 @@ Use cuGraph when the user's code is primarily:
|
||||
- Social network analysis, knowledge graphs, or recommendation systems
|
||||
- Any graph algorithm on networks with 10K+ edges
|
||||
|
||||
cuGraph's `nx-cugraph` backend can accelerate existing NetworkX code with zero code changes via an environment variable. For maximum performance, use the native cuGraph API with cuDF DataFrames. Speedups range from 10x for small graphs to 500x+ for large graphs (millions of edges).
|
||||
Start with the `nx-cugraph` backend and inspect fallback behavior. Move to the native cuGraph API
|
||||
with cuDF edge lists for unsupported operations or a measured performance reason. Include graph
|
||||
construction and host-device conversion in end-to-end benchmarks.
|
||||
|
||||
**Best for:** PageRank, betweenness centrality, community detection (Louvain, Leiden), BFS/SSSP, connected components, link prediction, graph neural network sampling, any NetworkX-heavy workflow.
|
||||
|
||||
@@ -108,12 +122,15 @@ KvikIO provides Python bindings to NVIDIA cuFile, enabling GPUDirect Storage (GD
|
||||
|
||||
**Note:** For tabular formats (CSV, Parquet, JSON), use cuDF's built-in readers instead — they're optimized for those formats. KvikIO is for raw binary data and remote file access.
|
||||
|
||||
### cuxfilter — for GPU-accelerated interactive dashboards
|
||||
### cuxfilter — legacy dashboards only
|
||||
**Read:** `references/cuxfilter.md`
|
||||
|
||||
**Project status: sunset.** RAPIDS 26.06 was cuxfilter's final release (RSN 60) — the packages still work but receive no further updates. For new dashboards, prefer cuDF for GPU data prep combined with HoloViews/hvPlot/Datashader linked selections, served with Panel, Plotly Dash, Streamlit, or Bokeh. Reach for cuxfilter only when the user already uses it or explicitly asks for it.
|
||||
**Project status: sunset.** RAPIDS 26.06 was cuxfilter's final release (RSN 60). Reach
|
||||
for cuxfilter only when the user already uses it or explicitly requests it. For new dashboards,
|
||||
prefer cuDF for GPU data prep combined with HoloViews/hvPlot/Datashader linked selections, served
|
||||
with Panel, Plotly Dash, Streamlit, or Bokeh.
|
||||
|
||||
Use cuxfilter when the user needs:
|
||||
Maintain cuxfilter when an existing application needs:
|
||||
- Interactive cross-filtering dashboards on large datasets (millions of rows)
|
||||
- Exploratory data analysis with linked charts that filter each other
|
||||
- GPU-accelerated visualization with scatter plots, bar charts, heatmaps, choropleths, or graph visualizations
|
||||
@@ -122,7 +139,8 @@ Use cuxfilter when the user needs:
|
||||
|
||||
cuxfilter leverages cuDF for all data operations on the GPU — filtering, groupby, and aggregation happen entirely on the GPU, with only rendering results sent to the browser. It integrates Bokeh, Datashader (for millions of points), Deck.gl (for maps), and Panel widgets.
|
||||
|
||||
**Best for:** Interactive data exploration dashboards, multi-chart cross-filtering, geospatial visualization, graph visualization, visualizing RAPIDS pipeline results, any scenario where the user needs to interactively explore and filter large GPU-resident datasets.
|
||||
**Best for:** Existing 26.06 applications that cannot yet migrate. Do not start a new dependency on
|
||||
an unmaintained dashboard framework.
|
||||
|
||||
### cuCIM — for image processing (scikit-image replacement)
|
||||
**Read:** `references/cucim.md`
|
||||
@@ -151,12 +169,12 @@ cuVS provides GPU-accelerated ANN index types (CAGRA, IVF-Flat, IVF-PQ, brute fo
|
||||
|
||||
**Best for:** Embedding search, RAG retrieval, recommender systems, image/text/audio similarity search, k-NN graph construction, any nearest-neighbor workload on 10K+ vectors.
|
||||
|
||||
### cuSpatial — for geospatial analytics (GeoPandas replacement)
|
||||
### cuSpatial — archived geospatial pipelines only
|
||||
**Read:** `references/cuspatial.md`
|
||||
|
||||
**Project status: archived.** The cuSpatial repository has been read-only since July 2025; the final release is 25.04, which pins `cudf-cu12==25.4.*` and therefore conflicts with current RAPIDS releases in the same environment. No official successor exists. Recommend it only in a dedicated legacy environment; otherwise keep geometry operations on GeoPandas/Shapely (CPU) and accelerate the tabular parts of the workflow with cuDF.
|
||||
|
||||
Use cuSpatial when the user's code is primarily:
|
||||
Maintain cuSpatial when an isolated 25.04 environment already uses:
|
||||
- GeoPandas spatial operations (point-in-polygon, spatial joins, distance calculations)
|
||||
- Trajectory analysis (grouping GPS traces, computing speeds/distances)
|
||||
- Spatial indexing (quadtree) for large-scale spatial joins
|
||||
@@ -165,7 +183,8 @@ Use cuSpatial when the user's code is primarily:
|
||||
|
||||
cuSpatial provides GPU-accelerated `GeoSeries` and `GeoDataFrame` types compatible with GeoPandas, plus spatial join, distance, and trajectory functions. Convert from GeoPandas with `cuspatial.from_geopandas()`.
|
||||
|
||||
**Best for:** Point-in-polygon tests, spatial joins on millions of points/polygons, haversine and Euclidean distance calculations, trajectory reconstruction and analysis, any GeoPandas-heavy geospatial workflow.
|
||||
**Best for:** Existing pipelines pinned to the 25.04 stack. Do not present it as a current
|
||||
GeoPandas replacement or combine it with current RAPIDS packages.
|
||||
|
||||
### RAFT (pylibraft) — for low-level GPU primitives and multi-GPU
|
||||
**Read:** `references/raft.md`
|
||||
@@ -200,16 +219,16 @@ Common combinations:
|
||||
- **CuPy + Numba**: Use CuPy for standard ops, drop into Numba for custom kernels
|
||||
- **cuDF + Numba**: Process dataframes with cuDF, apply custom GPU functions via Numba UDFs
|
||||
- **cuML + CuPy**: Train with cuML, do custom post-processing with CuPy
|
||||
- **cuDF + cuxfilter**: Load data with cuDF, build interactive cross-filtering dashboards with cuxfilter
|
||||
- **cuML + cuxfilter**: Run ML (e.g., UMAP, clustering) with cuML, visualize results interactively with cuxfilter
|
||||
- **cuGraph + cuxfilter**: Run graph analytics with cuGraph, visualize graph structure with cuxfilter's datashader graph chart
|
||||
- **cuDF + cuxfilter (legacy 26.06 only)**: Maintain an existing cross-filtering dashboard
|
||||
- **cuML + cuxfilter (legacy 26.06 only)**: Maintain an existing ML visualization workflow
|
||||
- **cuGraph + cuxfilter (legacy 26.06 only)**: Maintain an existing graph visualization
|
||||
- **cuCIM + CuPy**: cuCIM operates on CuPy arrays natively — chain image processing with array math
|
||||
- **cuCIM + PyTorch**: Preprocess images with cuCIM, pass directly to PyTorch via DLPack — zero-copy
|
||||
- **cuCIM + cuML**: Extract image features with cuCIM (regionprops), train classifiers with cuML
|
||||
- **KvikIO + CuPy**: Load raw binary data directly into CuPy arrays via GDS, bypassing CPU memory
|
||||
- **KvikIO + Numba**: Read data directly to GPU with KvikIO, process with custom Numba CUDA kernels
|
||||
- **KvikIO + Zarr**: Use GDSStore backend to read/write chunked N-dimensional arrays directly on GPU
|
||||
- **cuSpatial + cuDF**: Load geospatial data with cuDF, do spatial joins/analysis with cuSpatial
|
||||
- **cuSpatial + cuML**: Extract spatial features with cuSpatial, train ML models with cuML
|
||||
- **cuSpatial + cuDF (legacy 25.04 only)**: Keep both packages on the compatible frozen stack
|
||||
- **cuSpatial + cuML (legacy 25.04 only)**: Keep the full environment pinned and isolated
|
||||
- **RAFT + CuPy**: Use RAFT's eigsh() on sparse matrices built with CuPy/cupyx.scipy.sparse
|
||||
- **RAFT + raft-dask**: Scale GPU workloads across multiple GPUs/nodes via Dask
|
||||
|
||||
@@ -2,60 +2,60 @@
|
||||
|
||||
Per-library install commands, CUDA version selection, and environment setup.
|
||||
|
||||
## Installation
|
||||
Use `uv add` in standalone examples to match this repository's convention. If the user's project
|
||||
already uses another package manager, follow that project rather than rewriting its tooling.
|
||||
|
||||
IMPORTANT: Always use `uv add` for package installation — never `pip install` or `conda install`. This applies to install instructions in code comments, docstrings, error messages, and any other output you generate. If the user's project uses a different package manager, follow their lead, but default to `uv add`.
|
||||
|
||||
RAPIDS packages below track RAPIDS 26.06 (June 2026): they require Python >= 3.11 and CUDA 12.x or 13.x. Every RAPIDS package ships `-cu12` and `-cu13` wheel variants (except the archived cuSpatial) — the examples use `-cu12`; substitute `-cu13` for CUDA 13 systems. Most RAPIDS wheels are now published directly on PyPI; only cuGraph, nx-cugraph, and cuSpatial still require the NVIDIA index.
|
||||
RAPIDS packages below track RAPIDS 26.06 (June 2026): they require Python >= 3.11 and CUDA 12.x or 13.x. Every maintained RAPIDS package ships `-cu12` and `-cu13` wheel variants; the examples use `-cu12`, so substitute `-cu13` for CUDA 13 systems. Use the [RAPIDS release selector](https://docs.rapids.ai/install/) to confirm driver, Python, CUDA, and package compatibility. The NVIDIA extra index is included consistently because package availability differs; many packages are also mirrored directly on PyPI.
|
||||
|
||||
```bash
|
||||
# CuPy (choose the right CUDA version; CuPy 14+ supports CUDA 12/13 only)
|
||||
uv add cupy-cuda12x # For CUDA 12.x
|
||||
uv add cupy-cuda13x # For CUDA 13.x
|
||||
uv add "cupy-cuda12x==14.1.*" # For CUDA 12.x
|
||||
uv add "cupy-cuda13x==14.1.*" # For CUDA 13.x
|
||||
|
||||
# Numba with CUDA support (installs numba automatically)
|
||||
uv add "numba-cuda[cu12]" # or [cu13] — the NVIDIA package providing the numba.cuda target
|
||||
# Numba-CUDA compatibility path (maintenance mode; installs numba automatically)
|
||||
uv add "numba-cuda[cu12]==0.30.*" # use [cu13] for CUDA 13
|
||||
# For new kernel projects, evaluate Numba-CUDA-MLIR and its migration guide first.
|
||||
|
||||
# Warp (simulation, spatial computing, differentiable programming)
|
||||
uv add warp-lang # CUDA 12 runtime included; CUDA 13 builds are on GitHub Releases only
|
||||
uv add "warp-lang==1.15.*" # CUDA 12 runtime included; CUDA 13 builds are on GitHub Releases only
|
||||
|
||||
# cuDF (RAPIDS)
|
||||
uv add cudf-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cudf-cu12==26.6.*"
|
||||
# For cudf.pandas accelerator mode, that's all you need
|
||||
# Load it with: python -m cudf.pandas your_script.py
|
||||
|
||||
# cuML (RAPIDS machine learning)
|
||||
uv add cuml-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuml-cu12==26.6.*"
|
||||
# For cuml.accel accelerator mode (zero-change sklearn acceleration):
|
||||
# Load it with: python -m cuml.accel your_script.py
|
||||
|
||||
# cuGraph (RAPIDS graph analytics) — NVIDIA index still required (PyPI has only stub packages)
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cugraph-cu12 # Core cuGraph
|
||||
uv add --extra-index-url=https://pypi.nvidia.com nx-cugraph-cu12 # NetworkX backend
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cugraph-cu12==26.6.*" # Core cuGraph
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "nx-cugraph-cu12==26.6.*" # NetworkX backend
|
||||
# For nx-cugraph zero-change NetworkX acceleration:
|
||||
# NX_CUGRAPH_AUTOCONFIG=True python your_script.py
|
||||
|
||||
# KvikIO (high-performance GPU file IO)
|
||||
uv add kvikio-cu12
|
||||
# Optional: uv add zarr # For Zarr GPU backend support
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "kvikio-cu12==26.6.*"
|
||||
# Optional: uv add "zarr==3.*" # For Zarr GPU backend support
|
||||
|
||||
# cuxfilter (interactive dashboards) — SUNSET: 26.06 is the final release
|
||||
uv add cuxfilter-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuxfilter-cu12==26.6.*"
|
||||
# Depends on cuDF — installs it automatically
|
||||
|
||||
# cuCIM (RAPIDS image processing — scikit-image on GPU)
|
||||
uv add cucim-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cucim-cu12==26.6.*"
|
||||
|
||||
# cuVS (RAPIDS vector search)
|
||||
uv add cuvs-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuvs-cu12==26.6.*"
|
||||
|
||||
# cuSpatial (geospatial) — ARCHIVED: frozen at 25.04, pins cudf-cu12==25.4.*
|
||||
# Install only in a dedicated environment; NVIDIA index required
|
||||
uv add --extra-index-url=https://pypi.nvidia.com cuspatial-cu12
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "cuspatial-cu12==25.4.*"
|
||||
|
||||
# RAFT (low-level GPU primitives)
|
||||
uv add pylibraft-cu12 # Core primitives
|
||||
uv add raft-dask-cu12 # Multi-GPU support (optional)
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "pylibraft-cu12==26.6.*" # Core primitives
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "raft-dask-cu12==26.6.*" # Multi-GPU support (optional)
|
||||
```
|
||||
|
||||
To check CUDA availability after installation:
|
||||
@@ -95,6 +95,11 @@ print(kvikio.cufile_driver.get("is_gds_available")) # True if GDS is set up
|
||||
import cuxfilter
|
||||
print(cuxfilter.__version__) # Should print version
|
||||
|
||||
# cuCIM
|
||||
from cucim.skimage.filters import gaussian
|
||||
import cupy as cp
|
||||
print(gaussian(cp.zeros((8, 8), dtype=cp.float32), sigma=1).shape)
|
||||
|
||||
# cuVS
|
||||
from cuvs.neighbors import cagra
|
||||
import cupy as cp
|
||||
|
||||
@@ -24,13 +24,13 @@ KvikIO is part of the RAPIDS ecosystem and interoperates with CuPy, cuDF, Numba,
|
||||
|
||||
```bash
|
||||
# CUDA 12.x
|
||||
uv add kvikio-cu12
|
||||
uv add "kvikio-cu12==26.6.*"
|
||||
|
||||
# CUDA 13.x
|
||||
uv add kvikio-cu13
|
||||
uv add "kvikio-cu13==26.6.*"
|
||||
|
||||
# For Zarr support (optional)
|
||||
uv add zarr
|
||||
uv add "zarr==3.*"
|
||||
```
|
||||
|
||||
Verify installation:
|
||||
@@ -278,7 +278,7 @@ Zarr + KvikIO is useful for:
|
||||
- Bioinformatics (genomic arrays)
|
||||
- Any workload using chunked arrays that need GPU processing
|
||||
|
||||
Requires: `uv add zarr` in addition to kvikio.
|
||||
Requires: `uv add "zarr==3.*"` in addition to KvikIO 26.06.
|
||||
|
||||
---
|
||||
|
||||
|
||||
@@ -1,8 +1,12 @@
|
||||
# Numba CUDA Reference
|
||||
|
||||
Numba compiles Python directly into CUDA kernels, giving you full control over GPU threads, blocks, shared memory, and synchronization. Use Numba when your algorithm needs custom GPU logic that can't be expressed as standard array operations.
|
||||
The established Numba-CUDA target compiles Python into CUDA kernels with explicit control over
|
||||
threads, blocks, shared memory, and synchronization. It is now in maintenance mode. For new kernel
|
||||
projects, evaluate Numba-CUDA-MLIR first; use this reference for existing `numba.cuda` code,
|
||||
compatibility work, or features not yet available in the MLIR implementation.
|
||||
|
||||
> **Full documentation:** https://nvidia.github.io/numba-cuda/
|
||||
> **New-development path:** https://nvidia.github.io/numba-cuda-mlir/
|
||||
|
||||
## Table of Contents
|
||||
|
||||
@@ -29,14 +33,20 @@ Numba compiles Python directly into CUDA kernels, giving you full control over G
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
uv add "numba-cuda[cu12]" # For CUDA 12.x (pulls in numba and CUDA components)
|
||||
uv add "numba-cuda[cu13]" # For CUDA 13.x
|
||||
uv add "numba-cuda[cu12]==0.30.*" # For CUDA 12.x (pulls in numba and CUDA components)
|
||||
uv add "numba-cuda[cu13]==0.30.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
The NVIDIA `numba-cuda` package is the current implementation of the CUDA target (Numba's built-in target is deprecated). It implements functionality under the `numba.cuda` namespace — no code changes needed vs the old built-in target — and depends on `numba`, so a single install command suffices. Note: `numba-cuda` is now in maintenance mode (security and critical fixes through the CUDA 13 lifetime); NVIDIA's new feature development targets the separate `numba-cuda-mlir` package.
|
||||
The NVIDIA `numba-cuda` package is the out-of-tree implementation of the established
|
||||
`numba.cuda` target (Numba's built-in target is deprecated). It keeps the `numba.cuda` namespace
|
||||
and depends on `numba`, so a single install command suffices. NVIDIA limits this implementation to
|
||||
security and critical fixes through the CUDA 13 lifetime; new feature development targets the
|
||||
separate `numba-cuda-mlir` package. Follow its migration guide before choosing a compiler for new
|
||||
code because feature coverage continues to evolve.
|
||||
|
||||
**Requirements:** CUDA Toolkit 12 or 13. GPU with Compute Capability >= 5.0 (Maxwell or newer) on CUDA 12, or >= 7.5 (Turing or newer) on CUDA 13.
|
||||
|
||||
@@ -313,7 +323,9 @@ def kernel(points, distances):
|
||||
)
|
||||
```
|
||||
|
||||
**Cross-compilation note:** A function decorated with `@numba.jit` (CPU JIT) can also be called from CUDA kernels — useful for sharing logic between CPU and GPU code paths.
|
||||
CPU `@numba.jit` dispatchers are not CUDA device functions. If CPU and GPU paths need the same
|
||||
formula, keep a small undecorated source function and create explicit CPU (`@njit`) and GPU
|
||||
(`@cuda.jit(device=True)`) implementations or wrappers; test them against the same fixtures.
|
||||
|
||||
---
|
||||
|
||||
@@ -341,14 +353,21 @@ Multi-dimensional indexing works via tuples: `cuda.atomic.add(result, (row, col)
|
||||
|
||||
```python
|
||||
@cuda.jit
|
||||
def histogram(data, bins):
|
||||
def histogram(data, bins, min_value, max_value):
|
||||
i = cuda.grid(1)
|
||||
if i < data.size:
|
||||
bin_idx = int(data[i] * len(bins))
|
||||
if 0 <= bin_idx < len(bins):
|
||||
value = data[i]
|
||||
n_bins = bins.size
|
||||
if min_value <= value <= max_value:
|
||||
bin_idx = int((value - min_value) * n_bins / (max_value - min_value))
|
||||
if bin_idx == n_bins: # Include the rightmost edge.
|
||||
bin_idx = n_bins - 1
|
||||
cuda.atomic.add(bins, bin_idx, 1)
|
||||
```
|
||||
|
||||
Validate `max_value > min_value` on the host before launch. For production histograms, prefer
|
||||
CuPy/CUB unless a custom binning rule is required.
|
||||
|
||||
---
|
||||
|
||||
## GPU Ufuncs
|
||||
@@ -546,7 +565,8 @@ def matmul_shared(A, B, C):
|
||||
tx, ty = cuda.threadIdx.x, cuda.threadIdx.y
|
||||
|
||||
tmp = float32(0.0)
|
||||
for tile in range(cuda.gridDim.x):
|
||||
n_tiles = (A.shape[1] + TPB - 1) // TPB
|
||||
for tile in range(n_tiles):
|
||||
# Load tile into shared memory (with bounds check)
|
||||
col = tx + tile * TPB
|
||||
row = ty + tile * TPB
|
||||
@@ -563,11 +583,11 @@ def matmul_shared(A, B, C):
|
||||
C[y, x] = tmp
|
||||
```
|
||||
|
||||
### Parallel Prefix Sum (Scan)
|
||||
### Block-Local Inclusive Prefix Sum
|
||||
|
||||
```python
|
||||
@cuda.jit
|
||||
def inclusive_scan(data, output):
|
||||
def block_inclusive_scan(data, output):
|
||||
shared = cuda.shared.array(256, dtype=float32)
|
||||
tid = cuda.threadIdx.x
|
||||
i = cuda.grid(1)
|
||||
@@ -575,18 +595,26 @@ def inclusive_scan(data, output):
|
||||
shared[tid] = data[i] if i < data.size else 0
|
||||
cuda.syncthreads()
|
||||
|
||||
# Up-sweep
|
||||
# Hillis-Steele scan within one block. Two barriers prevent read/write races.
|
||||
offset = 1
|
||||
while offset < cuda.blockDim.x:
|
||||
addend = float32(0.0)
|
||||
if tid >= offset:
|
||||
shared[tid] += shared[tid - offset]
|
||||
offset *= 2
|
||||
addend = shared[tid - offset]
|
||||
cuda.syncthreads()
|
||||
if tid >= offset:
|
||||
shared[tid] += addend
|
||||
cuda.syncthreads()
|
||||
offset *= 2
|
||||
|
||||
if i < data.size:
|
||||
output[i] = shared[tid]
|
||||
```
|
||||
|
||||
This computes an independent scan per block, not a whole-array scan. A complete multi-block scan
|
||||
also scans block totals and adds block offsets. Prefer `cupy.cumsum()`/CUB unless a custom scan
|
||||
operator is required.
|
||||
|
||||
### Shared Memory Reduction
|
||||
|
||||
```python
|
||||
@@ -643,9 +671,11 @@ def stencil_1d(data, output, radius):
|
||||
|
||||
### GPU-Specific Tips
|
||||
|
||||
1. **Minimize host-device transfers.** Use `cuda.to_device()` and keep data on GPU across multiple kernel calls. Every PCI-e transfer is expensive (~12 GB/s) vs GPU memory bandwidth (~900+ GB/s).
|
||||
1. **Minimize host-device transfers.** Use `cuda.to_device()` and keep data on GPU across multiple
|
||||
kernel calls. Measure on the target system; interconnect and device-memory bandwidth vary widely.
|
||||
|
||||
2. **Use shared memory** for data reused across threads in a block. Shared memory bandwidth is ~10-100x higher than global memory.
|
||||
2. **Use shared memory** for data reused across threads in a block when profiling shows global
|
||||
memory traffic is limiting performance. Shared memory is finite and can reduce occupancy.
|
||||
|
||||
3. **Coalesce memory accesses.** Adjacent threads (consecutive `threadIdx.x`) should access adjacent memory locations. This lets the hardware combine accesses into fewer wide transactions.
|
||||
|
||||
@@ -653,7 +683,8 @@ def stencil_1d(data, output, radius):
|
||||
|
||||
5. **Use `fastmath=True`** when IEEE-754 strictness isn't required. Enables FMA, fast sqrt/division, and faster trig/exp/log for float32.
|
||||
|
||||
6. **Prefer float32 over float64** when precision allows. GPU float32 throughput is 2x-32x higher depending on the GPU (consumer GPUs heavily penalize float64).
|
||||
6. **Prefer float32 over float64** when precision and stability requirements allow. The throughput
|
||||
ratio is architecture-specific, so benchmark the target GPU.
|
||||
|
||||
7. **Use streams** to overlap data transfer with computation.
|
||||
|
||||
@@ -668,7 +699,8 @@ def stencil_1d(data, output, radius):
|
||||
- Don't use Python objects, strings, or dynamic memory allocation inside kernels — Numba CUDA supports a restricted Python subset.
|
||||
- Don't put `syncthreads()` inside divergent branches — if threads in a block take different paths through a barrier, behavior is undefined (deadlock or corruption).
|
||||
- Don't forget `cuda.synchronize()` before reading results on CPU — kernel launches are async.
|
||||
- Don't launch kernels with tiny data sizes — kernel launch overhead (~5-20us) dominates for small arrays.
|
||||
- Don't assume a custom kernel helps small arrays. Measure launch and transfer overhead on the
|
||||
target system and batch or fuse work when overhead dominates.
|
||||
|
||||
---
|
||||
|
||||
|
||||
@@ -21,16 +21,17 @@ RAFT (Reusable Accelerated Functions and Tools) is a RAPIDS library of GPU-accel
|
||||
|
||||
## Installation and Setup
|
||||
|
||||
Always use `uv add` (never `pip install` or `conda install`) in all install instructions, docstrings, comments, and error messages.
|
||||
Use `uv add` in standalone examples; follow the user's existing project package manager when one
|
||||
is already configured.
|
||||
|
||||
```bash
|
||||
# pylibraft (core library)
|
||||
uv add --extra-index-url=https://pypi.nvidia.com pylibraft-cu12 # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com pylibraft-cu13 # For CUDA 13.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "pylibraft-cu12==26.6.*" # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "pylibraft-cu13==26.6.*" # For CUDA 13.x
|
||||
|
||||
# raft-dask (multi-node multi-GPU support, optional)
|
||||
uv add --extra-index-url=https://pypi.nvidia.com raft-dask-cu12 # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com raft-dask-cu13 # For CUDA 13.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "raft-dask-cu12==26.6.*" # For CUDA 12.x
|
||||
uv add --extra-index-url=https://pypi.nvidia.com "raft-dask-cu13==26.6.*" # For CUDA 13.x
|
||||
```
|
||||
|
||||
pylibraft and raft-dask wheels (including the companion `libraft` wheel) are also published directly to PyPI, so the extra index is optional.
|
||||
|
||||
@@ -26,8 +26,8 @@ Unlike Numba CUDA (which gives you raw thread/block control) or CuPy (which repl
|
||||
## Installation
|
||||
|
||||
```bash
|
||||
uv add warp-lang # PyPI wheels built with the CUDA 12.9 runtime
|
||||
# uv add warp-lang[examples] # Includes USD and example dependencies
|
||||
uv add "warp-lang==1.15.*" # PyPI wheels built with the CUDA 12.9 runtime
|
||||
# uv add "warp-lang[examples]==1.15.*" # Includes USD and example dependencies
|
||||
```
|
||||
|
||||
Requires Python >= 3.10 and an NVIDIA driver >= 525 for the CUDA 12 wheels. CUDA 13.0 builds (driver >= 580) are published on the project's GitHub Releases page rather than PyPI.
|
||||
@@ -46,12 +46,13 @@ wp.init()
|
||||
|
||||
| Use Case | Best Choice | Why |
|
||||
|----------|------------|-----|
|
||||
| Physics simulation (particles, cloth, fluids) | **Warp** | Built-in spatial primitives, differentiable, simulation-oriented |
|
||||
| High-level rigid-body / robotics simulation | **Newton** | Maintained engine built on Warp; successor to removed `warp.sim` |
|
||||
| Custom physics kernels (particles, cloth, fluids) | **Warp** | Spatial primitives, autodiff, and explicit kernels |
|
||||
| Geometry processing (meshes, ray casting, SDFs) | **Warp** | Native mesh/volume/BVH types, spatial queries |
|
||||
| Differentiable simulation for ML training | **Warp** | Automatic forward/backward AD, PyTorch/JAX integration |
|
||||
| Robotics (kinematics, dynamics, control) | **Warp** | Transforms, quaternions, spatial vectors built-in |
|
||||
| NumPy array math (FFT, linear algebra, sorting) | **CuPy** | Drop-in NumPy replacement, wraps cuBLAS/cuFFT |
|
||||
| Custom CUDA kernels with raw thread control | **Numba** | Direct CUDA programming model, shared memory |
|
||||
| General custom CUDA kernels with explicit SIMT control | **Numba-CUDA-MLIR** or **Numba-CUDA** | Direct CUDA programming model and shared memory |
|
||||
| Data wrangling / ETL on tabular data | **cuDF** | pandas API on GPU |
|
||||
| ML training (sklearn-style) | **cuML** | scikit-learn API on GPU |
|
||||
|
||||
@@ -59,7 +60,9 @@ Warp and Numba both compile Python to CUDA, but serve different niches:
|
||||
- **Warp** excels at simulation/spatial workloads with its rich type system (vec3, quat, transform, mesh, volume) and automatic differentiation
|
||||
- **Numba** excels at raw CUDA programming where you need explicit thread/block control, shared memory management, and atomic operations on arbitrary data
|
||||
|
||||
Note: Warp's former ready-made physics engine module `warp.sim` was removed in Warp 1.10 — it has been superseded by the separate Newton library, which is built on Warp. Warp itself remains the tool for writing custom simulation kernels.
|
||||
Warp's former ready-made physics engine module `warp.sim` was removed in Warp 1.10. Use the
|
||||
separate Newton library, built on Warp, for maintained high-level simulation APIs. Warp itself
|
||||
remains the tool for writing custom kernels and domain primitives.
|
||||
|
||||
---
|
||||
|
||||
@@ -486,15 +489,19 @@ Keep data on GPU. Use `wp.array` on device, avoid `.numpy()` in inner loops.
|
||||
|
||||
### 3. Use Tile Operations for Reductions and GEMM
|
||||
|
||||
Tile-based reductions are 50x+ faster than per-thread atomics. Use `wp.tile()` + `wp.tile_sum()` + `wp.tile_atomic_add()` instead of `wp.atomic_add()`.
|
||||
Tile operations can reduce global traffic and atomic contention, but they also change resource use.
|
||||
Compare `wp.tile()` / `wp.tile_sum()` designs with the simpler atomic implementation on the target
|
||||
shape and GPU.
|
||||
|
||||
### 4. Prefer float32 Over float64
|
||||
|
||||
GPU float32 throughput is 2x-32x higher than float64.
|
||||
Use float32 only when the numerical contract permits it. The float32/float64 throughput ratio is
|
||||
architecture-specific; validate accuracy and benchmark the target GPU.
|
||||
|
||||
### 5. Kernel Caching
|
||||
|
||||
Warp caches compiled kernels between runs. First launch compiles (can take seconds); subsequent runs load from cache in milliseconds.
|
||||
Warp caches compiled kernels between runs. Exclude first-use compilation from steady-state kernel
|
||||
timing, but include it when one-shot application latency matters.
|
||||
|
||||
### 6. Object Lifetime
|
||||
|
||||
|
||||
Reference in New Issue
Block a user