From 0340bc6c68fea418d90a77dd06ebb0de1fe30291 Mon Sep 17 00:00:00 2001
From: atamazov <artem.tamazov@gmail.com>
Date: Fri, 29 Dec 2023 00:52:50 +0300
Subject: [PATCH 1/2] Automatically activate the new HIPRTC PCH adaptations
 starting from the 6.0.24000 version. Fix some build errors (#2465 + more)

(cherry picked from commit 4f695d975a2a6de2f167fc2925f3bad79fbaaf98)
---
 src/kernels/MIOpenLayerNorm.cpp                             | 6 ++++++
 src/kernels/miopen_cstdint.hpp                              | 2 +-
 src/kernels/miopen_type_traits.hpp                          | 2 +-
 .../include/utility/static_kernel_config.hpp                | 1 +
 4 files changed, 9 insertions(+), 2 deletions(-)

diff --git a/src/kernels/MIOpenLayerNorm.cpp b/src/kernels/MIOpenLayerNorm.cpp
index 98e1b56d56..74003de0f7 100644
--- a/src/kernels/MIOpenLayerNorm.cpp
+++ b/src/kernels/MIOpenLayerNorm.cpp
@@ -28,6 +28,12 @@
 #include <hip/hip_runtime.h>
 #endif
 
+#ifndef MIOPEN_DONT_USE_HIP_RUNTIME_HEADERS
+#include <hip/hip_fp16.h>
+#include <hip/hip_runtime.h>
+#endif
+
+#include "miopen_cstdint.hpp"
 #include "float_types.h"
 
 #if MIOPEN_USE_BFP16 == 1
diff --git a/src/kernels/miopen_cstdint.hpp b/src/kernels/miopen_cstdint.hpp
index 57d0d088b1..752805244b 100644
--- a/src/kernels/miopen_cstdint.hpp
+++ b/src/kernels/miopen_cstdint.hpp
@@ -30,7 +30,7 @@ typedef signed char int8_t;
 typedef unsigned char uint8_t;
 typedef signed short int16_t;
 typedef unsigned short uint16_t;
-#if HIP_PACKAGE_VERSION_FLAT >= 6001000000ULL
+#if HIP_PACKAGE_VERSION_FLAT >= 6000024000ULL
 typedef signed int int32_t;
 typedef unsigned int uint32_t;
 typedef __hip_internal::uint64_t uint64_t;
diff --git a/src/kernels/miopen_type_traits.hpp b/src/kernels/miopen_type_traits.hpp
index 352010e3ac..b9cf11cd63 100644
--- a/src/kernels/miopen_type_traits.hpp
+++ b/src/kernels/miopen_type_traits.hpp
@@ -76,7 +76,7 @@ struct remove_cv
     typedef typename remove_volatile<typename remove_const<T>::type>::type type;
 };
 
-#if HIP_PACKAGE_VERSION_FLAT >= 6001000000ULL
+#if HIP_PACKAGE_VERSION_FLAT >= 6000024000ULL
 template <class T, T v>
 struct integral_constant
 {
diff --git a/src/kernels/static_composable_kernel/include/utility/static_kernel_config.hpp b/src/kernels/static_composable_kernel/include/utility/static_kernel_config.hpp
index b4c9c2806d..d1a883b2ca 100644
--- a/src/kernels/static_composable_kernel/include/utility/static_kernel_config.hpp
+++ b/src/kernels/static_composable_kernel/include/utility/static_kernel_config.hpp
@@ -5,6 +5,7 @@
 #include "hip/hip_runtime.h"
 #include "hip/hip_fp16.h"
 #endif
+#include "miopen_cstdint.hpp"
 #include "bfloat16_dev.hpp"
 
 // index type: unsigned or signed

From 69a720d9100247883412f4329f55f51b9cc689ec Mon Sep 17 00:00:00 2001
From: atamazov <artem.tamazov@gmail.com>
Date: Fri, 29 Dec 2023 15:44:21 +0300
Subject: [PATCH 2/2] Remove duplicated includes.

---
 src/kernels/MIOpenLayerNorm.cpp | 5 -----
 1 file changed, 5 deletions(-)

diff --git a/src/kernels/MIOpenLayerNorm.cpp b/src/kernels/MIOpenLayerNorm.cpp
index 74003de0f7..b73632b9c5 100644
--- a/src/kernels/MIOpenLayerNorm.cpp
+++ b/src/kernels/MIOpenLayerNorm.cpp
@@ -28,11 +28,6 @@
 #include <hip/hip_runtime.h>
 #endif
 
-#ifndef MIOPEN_DONT_USE_HIP_RUNTIME_HEADERS
-#include <hip/hip_fp16.h>
-#include <hip/hip_runtime.h>
-#endif
-
 #include "miopen_cstdint.hpp"
 #include "float_types.h"