oneapi-src / SYCLomatic

Other
222 stars 91 forks source link

[SYCLomatic] Fix incorrect typecast of `sycl::malloc` return type in case of CUDA type redefination #2146

Open the-slow-one opened 1 month ago

the-slow-one commented 1 month ago

The following migrated code in DPCT have incompatiable cast operation. This isssue is only relavent on Windows. The sample below only compiles on Windows.

// issue.dp.cpp
#include <sycl/sycl.hpp>
#include <dpct/dpct.hpp>
#include <cstdint>

typedef uint64_t CUdeviceptr;

void foo(dpct::device_ptr ptr) {
  ptr = DPCT_CHECK_ERROR(
      DPCT_CHECK_ERROR(ptr = (unsigned long long)sycl::malloc_device(
                           1024, dpct::get_in_order_queue())));
}

Orginal CUDA code is

#include <cstdint>
#include <cuda.h>

typedef uint64_t CUdeviceptr;

void foo(CUdeviceptr ptr) {
  ptr = cuMemAlloc(&ptr, 1024);
}

Solution is to stop the underlying type lookup if the type is a defined by CUDA. (Existing implementation only looks if the source of defination is cuda.h but it's not sufficent as evident in the CUDA example above)

zhiweij1 commented 12 hours ago

@the-slow-one The test case has syntax error, it cannot be compiled by nvcc:

zhiwei@zhiwei-ubuntu:~/testfolder$ nvcc test.cu -c
test.cu(4): error: invalid redeclaration of type name "CUdeviceptr" (declared at line 262 of /usr/local/cuda-12.4/bin/../targets/x86_64-linux/include/cuda.h)
  typedef uint64_t CUdeviceptr;
                   ^

test.cu(7): error: argument of type "CUdeviceptr *" is incompatible with parameter of type "CUdeviceptr *"
    ptr = cuMemAlloc_v2(&ptr, 1024);
                        ^

2 errors detected in the compilation of "test.cu".
zhiwei@zhiwei-ubuntu:~/testfolder$ 
the-slow-one commented 7 hours ago

@the-slow-one The test case has syntax error, it cannot be compiled by nvcc:

zhiwei@zhiwei-ubuntu:~/testfolder$ nvcc test.cu -c
test.cu(4): error: invalid redeclaration of type name "CUdeviceptr" (declared at line 262 of /usr/local/cuda-12.4/bin/../targets/x86_64-linux/include/cuda.h)
  typedef uint64_t CUdeviceptr;
                   ^

test.cu(7): error: argument of type "CUdeviceptr *" is incompatible with parameter of type "CUdeviceptr *"
    ptr = cuMemAlloc_v2(&ptr, 1024);
                        ^

2 errors detected in the compilation of "test.cu".
zhiwei@zhiwei-ubuntu:~/testfolder$ 

This is a Windows-only test. nvcc compiles this file on Windows.