From 824b81c03441d8df2d4f2ffc9f3ac692c0e159a6 Mon Sep 17 00:00:00 2001
From: corey-derochie-amd
 <161367113+corey-derochie-amd@users.noreply.github.com>
Date: Thu, 13 Feb 2025 10:00:22 -0700
Subject: [PATCH] =?UTF-8?q?Revert=20"replacing=20rccl=5Ffloat8=20with=20hi?=
 =?UTF-8?q?p=5Ffp8=20and=20address=20compatibility=20issue=20(#=E2=80=A6"?=
 =?UTF-8?q?=20(#1545)?=
MIME-Version: 1.0
Content-Type: text/plain; charset=UTF-8
Content-Transfer-Encoding: 8bit

This reverts commit d437d6e41c0af3592bca2363dc3b5ee418f6c5db.
---
 src/include/rccl_float8.h | 51 +++------------------------------------
 1 file changed, 3 insertions(+), 48 deletions(-)

diff --git a/src/include/rccl_float8.h b/src/include/rccl_float8.h
index 64c1b137e8..01cab41f71 100644
--- a/src/include/rccl_float8.h
+++ b/src/include/rccl_float8.h
@@ -25,7 +25,7 @@
 
 #include <stdint.h>
 
-#if __cplusplus < 201103L || (!defined(__HIP_PLATFORM_AMD__) && !defined(__HIPCC__))
+#if __cplusplus < 201103L || (!defined(__HCC__) && !defined(__HIPCC__))
 /*! \brief Struct to represent a 8 bit floating-point number. */
 
 typedef struct
@@ -38,52 +38,7 @@ typedef struct
     uint8_t data;
 } rccl_bfloat8;
 
-// __cplusplus < 201103L || (!defined(__HIP_PLATFORM_AMD__) && !defined(__HIPCC__))
-#elif ROCM_VERSION >= 60200
-
-#include <hip/hip_fp8.h>
-
-#if HIP_FP8_TYPE_OCP
-typedef __hip_fp8_e4m3 rccl_float8;
-typedef __hip_fp8_e5m2 rccl_bfloat8;
-#else
-typedef __hip_fp8_e4m3_fnuz rccl_float8;
-typedef __hip_fp8_e5m2_fnuz rccl_bfloat8;
-#endif
-    
-inline std::ostream& operator<<(std::ostream& os, const rccl_float8& f8)
-{
-    return os << float(f8);
-}
-
-inline std::ostream& operator<<(std::ostream& os, const rccl_bfloat8& bf8)
-{
-    return os << float(bf8);
-}
-
-inline __host__ __device__ float operator*(rccl_float8 a, rccl_float8 b)
-{
-    return float(a) * float(b);
-}
-
-inline __host__ __device__ float operator*(rccl_bfloat8 a, rccl_bfloat8 b)
-{
-    return float(a) * float(b);
-}
-
-inline __host__ __device__ float operator*(rccl_float8 a, float b)
-{
-    return float(a) * float(b);
-}
-
-inline __host__ __device__ float operator*(rccl_bfloat8 a, float b)
-{
-    return float(a) * float(b);
-}
-
-// For older versions of ROCm that do not include hip_fp8.h,
-// we provide a local version of the header file as a fallback.
-#else
+#else // __cplusplus < 201103L || (!defined(__HCC__) && !defined(__HIPCC__))
 
 #define HIP_HOST_DEVICE __host__ __device__
 #define HIP_HOST __host__
@@ -1061,6 +1016,6 @@ inline __host__ __device__ T explicit_downcast(Ta a, uint32_t rng)
 
 // =================================================================================================
 
-#endif
+#endif // __cplusplus < 201103L || (!defined(__HCC__) && !defined(__HIPCC__))
 
 #endif // ROCBLAS_FLOAT8_H