CUDA 13.0 SM_120 Support

[Arthav24]

Currently only x86_64_cuda12.8 version is available for libspconv. Running the BEV Fusion on Cuda 13.0 throws errors related to spconv. Pasting few below

[implicit-gemm.cu:303] (Error) CUDA Runtime error init_quantization_scale<FeaturesType::INT8><<<grid, block, 0, _stream>>>( input_scale, output_scale, weight_, prod_DHWI, int8weight_.ptr(), weight_scales_.ptr(), alpha_.ptr(), (half *)bias, bias_scaled_.ptr(), numel) # no kernel image is available for execution on the device, code = cudaErrorNoKernelImageForDevice [ 209 ] in file src/spconv/implicit-gemm.cu:303
[implicit-gemm.cu:307] (Error) CUDA Runtime error init_quantization_scale<FeaturesType::FP16><<<grid, block, 0, _stream>>>( input_scale, output_scale, weight_, prod_DHWI, int8weight_.ptr(), weight_scales_.ptr(), alpha_.ptr(), (half *)bias, bias_scaled_.ptr(), numel) # no kernel image is available for execution on the device, code = cudaErrorNoKernelImageForDevice [ 209 ] in file src/spconv/implicit-gemm.cu:307
[rulebook.cu:273] (Error) CUDA Runtime error spconv::setup_hash_and_route<<<(num_input + 1023) / 1024, 1024, 0, stream>>>( hash_.get(), route_mask_.ptr(), route_.ptr(), indices, num_input, prob->kv) # no kernel image is available for execution on the device, code = cudaErrorNoKernelImageForDevice [ 209 ] in file src/spconv/rulebook.cu:273
[rulebook.cu:278] (Error) CUDA Runtime error spconv::build_subm_rulebook<<<grid, block, 0, stream>>>( indices, route_mask_.ptr(), route_.ptr(), num_input, *prob, this->hash_.get()) # no kernel image is available for execution on the device, code = cudaErrorNoKernelImageForDevice [ 209 ] in file src/spconv/rulebook.cu:278
[implicit-gemm.cu:191] (Error) CUDA Runtime error gemmfn<<<grid, block, memsize * num_warp, stream>>>( book->route(), book->route_mask(), book->relocation(), (half *)bias_, fp16_input_ptr, int8output_.ptr(), (half *)weight_, prob.Ci, M, nullptr, N, K, prob.kv, output_scale_, true) # no kernel image is available for execution on the device, code = cudaErrorNoKernelImageForDevice [ 209 ] in file src/spconv/implicit-gemm.cu:191
[engine.cu:1357] (Error) CUDA Runtime error add_and_relu_kernel_quant<<<grid_size, block_size, 0, stream>>>( a.features.ptr<signed char>(), ascale, b.features.ptr<signed char>(), bscale, c.features.ptr<signed char>(), 1.0f/cscale, n) # no kernel image is available for execution on the device, code = cudaErrorNoKernelImageForDevice [ 209 ] in file src/spconv/engine.cu:1357

Are there any plans to release spconv for cuda 13.0 on x86_64 or is there any way to build this spconv by source for cuda 13 and sm120 as there is support for cuda 13 on aarch64. I tried building https://github.com/traveller59/spconv and running example. it throws an error.

[Arthav24]

I tried on sm_75 RTX2060 with cuda 12.8 got same error.

[hopef]

Currently, libspconv only supports CUDA 11.4 and 12.8 on platforms with SM_80 or higher.

[Arthav24]

I tried on sm_120 (RTX5070Ti) initially. But spconv throwed errors

[hopef]

Please try using the newest version that I uploaded 3 days ago.

[Arthav24]

Hey @hopef Thank you so much for the update. I just tested it on my system it works now. Thanks again