cusparseSpGEMM accessing uninitilized memory?

[JiamingCao]

When I try the spgemm_example.c example, it does run correctly, but curiously, if I pass it through the initcheck tool of compute-sanitizer, I get errors stating that cusparseSpGEMM_compute accesses uninitialized global memory. Could this be a bug in CUDA, or am I missing something silly?

My system:

• Ubuntu 22.04.1, kernel 6.8.0
• GPU: GTX 980Ti
• Driver: 550.78
• GCC: 11.4.0
• CUDA toolkit (installed through the .run file): V12.4.131

Command I used to compile:

nvcc -O3 spgemm_example.c -lcusparse -o spgemm_example

compute-sanitizer gives:

========= COMPUTE-SANITIZER
========= Uninitialized __global__ memory read of size 4 bytes
=========     at void cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::DeviceScanKernel<cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::DeviceScanPolicy<int>::Policy600, const int *, int *, cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::ScanTileState<int, (bool)1>, cuda::std::__4::plus<void>, cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::detail::InputValue<int, int *>, int, int>(T2, T3, T4, int, T5, T6, T7)+0x2e8
=========     by thread (0,0,0) in block (0,0,0)
=========     Address 0xb065c1678
=========     Saved host backtrace up to driver entry point at kernel launch time
=========     Host Frame: [0x2efadf]
=========                in /lib/x86_64-linux-gnu/libcuda.so.1
=========     Host Frame: [0x9bf16d]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0xa228ad]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x89a1b6]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6f3c04]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6d4b9c]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6ebf91]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame:cusparseSpGEMM_compute [0xf5624]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame:main [0xb011]
=========                in /home/jiaming/Documents/UoBResearch/Research/nirfastff/NIRFASTerPy/testing/./spgemm_example
=========     Host Frame:__libc_start_call_main in ../sysdeps/nptl/libc_start_call_main.h:58 [0x29d8f]
=========                in /lib/x86_64-linux-gnu/libc.so.6
=========     Host Frame:__libc_start_main in ../csu/libc-start.c:392 [0x29e3f]
=========                in /lib/x86_64-linux-gnu/libc.so.6
=========     Host Frame:_start [0xb9e4]
=========                in /home/jiaming/Documents/UoBResearch/Research/nirfastff/NIRFASTerPy/testing/./spgemm_example
========= 
========= Uninitialized __global__ memory read of size 4 bytes
=========     at void cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::DeviceScanKernel<cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::DeviceScanPolicy<int>::Policy600, const int *, int *, cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::ScanTileState<int, (bool)1>, cuda::std::__4::plus<void>, cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::detail::InputValue<int, int *>, int, int>(T2, T3, T4, int, T5, T6, T7)+0x2e8
=========     by thread (0,0,0) in block (0,0,0)
=========     Address 0xb065c1498
=========     Saved host backtrace up to driver entry point at kernel launch time
=========     Host Frame: [0x2efadf]
=========                in /lib/x86_64-linux-gnu/libcuda.so.1
=========     Host Frame: [0x9bf16d]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0xa228ad]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x89a1b6]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6f3c04]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6d4f57]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6ebf91]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame:cusparseSpGEMM_compute [0xf5624]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame:main [0xb011]
=========                in /home/jiaming/Documents/UoBResearch/Research/nirfastff/NIRFASTerPy/testing/./spgemm_example
=========     Host Frame:__libc_start_call_main in ../sysdeps/nptl/libc_start_call_main.h:58 [0x29d8f]
=========                in /lib/x86_64-linux-gnu/libc.so.6
=========     Host Frame:__libc_start_main in ../csu/libc-start.c:392 [0x29e3f]
=========                in /lib/x86_64-linux-gnu/libc.so.6
=========     Host Frame:_start [0xb9e4]
=========                in /home/jiaming/Documents/UoBResearch/Research/nirfastff/NIRFASTerPy/testing/./spgemm_example
========= 
========= Uninitialized __global__ memory read of size 4 bytes
=========     at void cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::DeviceScanKernel<cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::DeviceScanPolicy<int>::Policy600, const int *, int *, cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::ScanTileState<int, (bool)1>, cuda::std::__4::plus<void>, cub::CUB_200101_500_520_600_610_700_750_800_860_890_900_NS::detail::InputValue<int, int *>, int, int>(T2, T3, T4, int, T5, T6, T7)+0x2e8
=========     by thread (0,0,0) in block (0,0,0)
=========     Address 0xb065c14a8
=========     Saved host backtrace up to driver entry point at kernel launch time
=========     Host Frame: [0x2efadf]
=========                in /lib/x86_64-linux-gnu/libcuda.so.1
=========     Host Frame: [0x9bf16d]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0xa228ad]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x89a1b6]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6f3c04]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6d503f]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame: [0x6ebf91]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame:cusparseSpGEMM_compute [0xf5624]
=========                in /usr/local/cuda-12.4/lib64/libcusparse.so.12
=========     Host Frame:main [0xb011]
=========                in /home/jiaming/Documents/UoBResearch/Research/nirfastff/NIRFASTerPy/testing/./spgemm_example
=========     Host Frame:__libc_start_call_main in ../sysdeps/nptl/libc_start_call_main.h:58 [0x29d8f]
=========                in /lib/x86_64-linux-gnu/libc.so.6
=========     Host Frame:__libc_start_main in ../csu/libc-start.c:392 [0x29e3f]
=========                in /lib/x86_64-linux-gnu/libc.so.6
=========     Host Frame:_start [0xb9e4]
=========                in /home/jiaming/Documents/UoBResearch/Research/nirfastff/NIRFASTerPy/testing/./spgemm_example
========= 
spgemm_example test PASSED
========= ERROR SUMMARY: 3 errors