fix 64 bit int types in type_dispatcher.cuh
@@ -40,11 +40,11 @@ void dtype_dispatcher(int dtype_id, functor_t f, Ts&&... args)
case CUPY_TYPE_INT8: return f.template operator()<char>(std::forward<Ts>(args)...);
case CUPY_TYPE_INT16: return f.template operator()<short>(std::forward<Ts>(args)...);
case CUPY_TYPE_INT32: return f.template operator()<int>(std::forward<Ts>(args)...);
- case CUPY_TYPE_INT64: return f.template operator()<long>(std::forward<Ts>(args)...);
+ case CUPY_TYPE_INT64: return f.template operator()<int64_t>(std::forward<Ts>(args)...);
case CUPY_TYPE_UINT8: return f.template operator()<unsigned char>(std::forward<Ts>(args)...);
case CUPY_TYPE_UINT16: return f.template operator()<unsigned short>(std::forward<Ts>(args)...);
case CUPY_TYPE_UINT32: return f.template operator()<unsigned int>(std::forward<Ts>(args)...);
- case CUPY_TYPE_UINT64: return f.template operator()<unsigned long>(std::forward<Ts>(args)...);
+ case CUPY_TYPE_UINT64: return f.template operator()<uint64_t>(std::forward<Ts>(args)...);
#if ((__CUDACC_VER_MAJOR__ > 9 || (__CUDACC_VER_MAJOR__ == 9 && __CUDACC_VER_MINOR__ == 2)) \
&& (__CUDA_ARCH__ >= 530 || !defined(__CUDA_ARCH__))) || (defined(__HIPCC__) || defined(CUPY_USE_HIP))
case CUPY_TYPE_FLOAT16: return f.template operator()<__half>(std::forward<Ts>(args)...);
