pytorch_scatter icon indicating copy to clipboard operation
pytorch_scatter copied to clipboard

scatter2.0.9 no matching function for call to '__ldg'

Open zhong110020 opened this issue 1 year ago • 4 comments

csrc/hip/segment_csr_cuda.hip:188:20: error: no matching function for call to '__ldg' scalar_t val = __ldg(src_data + row_idx); ^~~~~ csrc/hip/segment_csr_cuda.hip:275:27: note: in instantiation of function template specialization 'gather_csr_kernel<c10::Half, 4>' requested here hipLaunchKernelGGL(( gather_csr_kernel<scalar_t, 4>), dim3(BLOCKS(1, 4 * N)), dim3(THREADS), 0, stream,

zhong110020 avatar Jan 26 '24 09:01 zhong110020

Can you share some more information about your system and PyTorch version? Can you post the full installation log?

rusty1s avatar Jan 28 '24 14:01 rusty1s

Can you share some more information about your system and PyTorch version? Can you post the full installation log?

pytorch11.8,use pytorch_scatter0.2.9version,add environment variables export FORCE_CUDA=1,Compile Command python3 setup.py install --user bdist_wheel,The compilation process reported an error as follows: ture date. Please use a non-deprecated interface with equivalent functionality instead. For a listing of replacement headers and interfaces, consult the file backward_warning.h. To disable this warning use -Wno-deprecated. [-W#warnings] #warning
^ csrc/hip/segment_csr_cuda.hip:108:3: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result] hipSetDevice(src.get_device()); ^~~~~~~~~~~~ ~~~~~~~~~~~~~~~~ csrc/hip/segment_csr_cuda.hip:228:3: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result] hipSetDevice(src.get_device()); ^~~~~~~~~~~~ ~~~~~~~~~~~~~~~~ csrc/hip/segment_csr_cuda.hip:188:20: error: no matching function for call to '__ldg' scalar_t val = __ldg(src_data + row_idx); ^~~~~ csrc/hip/segment_csr_cuda.hip:275:27: note: in instantiation of function template specialization 'gather_csr_kernel<c10::Half, 4>' requested here hipLaunchKernelGGL(( gather_csr_kernel<scalar_t, 4>), dim3(BLOCKS(1, 4 * N)), dim3(THREADS), 0, stream, ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:30:31: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const char ' for 1st argument device inline static char __ldg(const char ptr) { return *ptr; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:32:32: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const char2 *' (aka 'const HIP_vector_type<char, 2> ') for 1st argument device inline static char2 __ldg(const char2 ptr) { return *ptr; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:34:32: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const char4 *' (aka 'const HIP_vector_type<char, 4> ') for 1st argument device inline static char4 __ldg(const char4 ptr) { return *ptr; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:36:38: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const signed char ' for 1st argument device inline static signed char __ldg(const signed char ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:38:40: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const unsigned char ' for 1st argument device inline static unsigned char __ldg(const unsigned char ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:41:32: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const short ' for 1st argument device inline static short __ldg(const short ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:43:33: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const short2 *' (aka 'const HIP_vector_type<short, 2> ') for 1st argument device inline static short2 __ldg(const short2 ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:45:33: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const short4 *' (aka 'const HIP_vector_type<short, 4> ') for 1st argument device inline static short4 __ldg(const short4 ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:47:41: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const unsigned short ' for 1st argument device inline static unsigned short __ldg(const unsigned short ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:50:30: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const int ' for 1st argument device inline static int __ldg(const int ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:52:31: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const int2 *' (aka 'const HIP_vector_type<int, 2> ') for 1st argument device inline static int2 __ldg(const int2 ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:54:31: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const int4 *' (aka 'const HIP_vector_type<int, 4> ') for 1st argument device inline static int4 __ldg(const int4 ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:56:39: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const unsigned int ' for 1st argument device inline static unsigned int __ldg(const unsigned int ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:59:31: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const long ' for 1st argument device inline static long __ldg(const long ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:61:40: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const unsigned long ' for 1st argument device inline static unsigned long __ldg(const unsigned long ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:64:36: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const long long ' for 1st argument device inline static long long __ldg(const long long ptr) { return ptr[0]; } ^ /opt/dtk/hip/include/hip/amd_detail/hip_ldg.h:66:36: note: candidate function not viable: no known conversion from 'const c10::Half *' to 'const longlong2 *' (aka 'const HIP_vector_type<long long, 2> ') for 1st argument device inline static longlong2 __ldg(const longlong2 ptr) { return ptr[0]; }

zhong110020 avatar Jan 29 '24 00:01 zhong110020

cuda11.8,pytorch1.13

zhong110020 avatar Jan 29 '24 01:01 zhong110020

Can you try if running

pip install torch-scatter==2.0.9 --no-index -f https://data.pyg.org/whl/torch-1.13.0+cu117.html

fixes your issues?

rusty1s avatar Jan 29 '24 06:01 rusty1s

This issue had no activity for 6 months. It will be closed in 2 weeks unless there is some new activity. Is this issue already resolved?

github-actions[bot] avatar Jul 28 '24 01:07 github-actions[bot]