When working with large datasets, illegal memory access errors often arise because of truncated integer ranges. Catching sign changes as well as signed-vs-unsigned comparisons could help prevent this.
Ideally, I'd like to use something akin to LLVM's -Wsign-conversion
and -Wsign-compare
to flag stop compilation when an integer's signedness is implicitly changed or signed and unsigned integers are compared.
Looking through this list I see integer_sign_change
. Perhaps that is akin to -Wsign-conversion
? I don't see anything for flagging signed comparisons.
When I compile this program
#include <cstdint>
#include <limits>
#include <string>
__global__ void foo(uint *d, uint N) {
int a = 3;
unsigned int b = 4;
a = b; // Expecting a sign conversion warning
b = a; // Expecting a sign conversion warning
const int tid = threadIdx.x + blockIdx.x * blockDim.x;
// Expecting a sign comparison warning
for(int i = tid; i < N; i += blockDim.x * gridDim.x){
}
}
int main(int argc, char **argv){
uint *d = nullptr;
if (argc != 2){ return -1; }
int N = std::stoi(argv[1]);
// Arbitrary launch parameters
foo<<<1, 1>>>(d, N);
}
following this advice to use, eg:
nvcc test.cu -Xcudafe="--diag_error=integer_sign_change"
there is no compilation error.
Interestingly, if I pass similar code through our AMD/HIP/Rocm stack, using -Werror,-Wsign-compare
does produce a handy compilation error.
Is there a way to have NVCC throw errors on sign conversion and sign comparisons?