Illegal memory access on Blackwell (RTX 5090) — fix available upstream #2
Description
Problem
CuMesh causes CUDA error: illegal memory access on NVIDIA Blackwell GPUs (RTX 5090, sm_120) when used with PyTorch and ComfyUI-Trellis2.
Setting CUDA_LAUNCH_BLOCKING=1 masks the issue, confirming it is a stream synchronization bug.
Root cause
All CUDA kernel launches and CUB operations use the default stream (stream 0). PyTorch uses cudaStreamNonBlocking streams, which have no implicit synchronization with stream 0. This is undefined behavior per the CUDA programming model but only manifests reliably on Blackwell GPUs.
Fix
A comprehensive fix has been submitted upstream and is available here:
- Upstream PR: Fix Blackwell (sm_120) illegal memory access via CUDA stream handling JeffreyXiang/CuMesh#29
- Branch with fix for this repo: https://github.com/cuzelac/CuMesh/tree/fix/blackwell-stream-sync-v2
The fix passes at::cuda::getCurrentCUDAStream() to all kernel launches, CUB calls, and adds cudaStreamSynchronize(stream) before cudaFree operations. This follows the same pattern used by nvdiffrast and cubvh.
Testing
- Tested on RTX 5090 (sm_120), PyTorch 2.10.0+cu130, CUDA 13.0, Windows
- Full Trellis2 image-to-3D pipeline including mesh generation, refinement, and texturing
- Multiple successful runs with zero crashes
I was unable to submit a PR directly to this repo due to GitHub fork network limitations — happy to help if you'd like to pull the changes in.