NVIDIA / NVIDIA/CUDALibrarySamples

cusparseSpGEMM accessing uninitilized memory?

Open
#225 3 comments 0 reactions 1 assignee View on GitHub

@essex-edwards is already working on this.

Since Jun 27, 2025.

cuSPARSE
Dominant language
Cuda
Stars
2.5k
Forks
478
PR merge metrics
No merged PRs in 30d

Description

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

Contributor guide

Open the contributing guide

First steps

  1. Read the whole issue, then the project's contributing guide.
  2. Comment on the issue to say you are picking it up — it saves two people doing the same work.
  3. Fork the repository and make your change on a branch.
  4. Open a pull request that references the issue number.

Assessment

This issue has not been assessed yet.

Get new issues in your inbox

A short digest of beginner-friendly GitHub issues.