最新消息:雨落星辰是一个专注网站SEO优化、网站SEO诊断、搜索引擎研究、网络营销推广、网站策划运营及站长类的自媒体原创博客

How can I use NVCC to detect signedunsigned issues in CUDA kernel code? - Stack Overflow

programmeradmin6浏览0评论

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?

发布评论

评论列表(0)

  1. 暂无评论