From 6dcd2642aae67d0bc46eee959a9f196e6ea51e7c Mon Sep 17 00:00:00 2001 From: Jiandong Ruan Date: Sat, 8 Sep 2018 12:05:33 -0700 Subject: [PATCH] bug #1526 - CUDA compilation fails on CUDA 9.x SDK when arch is set to compute_60 and/or above --- Eigen/src/Core/arch/GPU/Half.h | 8 ++++++++ 1 file changed, 8 insertions(+) diff --git a/Eigen/src/Core/arch/GPU/Half.h b/Eigen/src/Core/arch/GPU/Half.h index b44a4356a..aef1eb079 100644 --- a/Eigen/src/Core/arch/GPU/Half.h +++ b/Eigen/src/Core/arch/GPU/Half.h @@ -272,7 +272,11 @@ namespace half_impl { // conversion steps back and forth. EIGEN_STRONG_INLINE __device__ half operator + (const half& a, const half& b) { +#if defined(EIGEN_CUDACC_VER) && EIGEN_CUDACC_VER >= 90000 + return __hadd(::__half(a), ::__half(b)); +#else return __hadd(a, b); +#endif } EIGEN_STRONG_INLINE __device__ half operator * (const half& a, const half& b) { return __hmul(a, b); @@ -281,9 +285,13 @@ EIGEN_STRONG_INLINE __device__ half operator - (const half& a, const half& b) { return __hsub(a, b); } EIGEN_STRONG_INLINE __device__ half operator / (const half& a, const half& b) { +#if defined(EIGEN_CUDACC_VER) && EIGEN_CUDACC_VER >= 90000 + return __hdiv(a, b); +#else float num = __half2float(a); float denom = __half2float(b); return __float2half(num / denom); +#endif } EIGEN_STRONG_INLINE __device__ half operator - (const half& a) { return __hneg(a);