[COMGR] Use OpenCL 2.0. [HIPRTC] Provide min/max limits for int. Fix …

…build errors related to min/max limits for BF16. (#2705)

* fix-rocm-mainline-issues-01(01) Removeed `constexpr` from numeric_limits<hip_bfloat16>::min()/max() as BF16 ctor provided by HIP can't be used in const expressions.

* fix-rocm-mainline-issues-01(02) [COMGR] Globally engage OpenCL 2.0

* fix-rocm-mainline-issues-01(03) [HIPRTC] Provide min/max limits for int

src/comgr.cpp

@@ -189,14 +189,17 @@ namespace ocl {
 
 #define OCL_EARLY_INLINE 1
 
-#define OCL_STANDARD 120 // For experiments.
+#define OCL_STANDARD 200
 
 #if !(OCL_STANDARD == 200 || OCL_STANDARD == 120)
 #error "Wrong OCL_STANDARD"
 #endif
 
 static void AddCompilerOptions(OptionList& list)
 {
+#if OCL_STANDARD == 200
+    list.push_back("-cl-std=CL2.0");
+#endif
     list.push_back("-cl-kernel-arg-info");
 #if 0 // For experimients.
     list.push_back("-cl-denorms-are-zero");

src/kernels/miopen_limits.hpp

@@ -66,29 +66,29 @@ template <>
 class numeric_limits<hip_bfloat16>
 {
 public:
-    static
-#if HIP_PACKAGE_VERSION_FLAT >= 6000000000ULL
-        constexpr
-#endif
-        __device__ hip_bfloat16
-        max() noexcept
+    static __device__ hip_bfloat16 max() noexcept
     {
         // data = 0x7F7F
         return static_cast<hip_bfloat16>(0x1.FEp+127f);
     }
 
-    static
-#if HIP_PACKAGE_VERSION_FLAT >= 6000000000ULL
-        constexpr
-#endif
-        __device__ hip_bfloat16
-        min() noexcept
+    static __device__ hip_bfloat16 min() noexcept
     {
         // data = 0x0080
         return static_cast<hip_bfloat16>(0x1p-14f);
     }
 };
 
+#if HIP_PACKAGE_VERSION_FLAT >= 6001024024ULL
+template <>
+class numeric_limits<int>
+{
+public:
+    static constexpr __device__ int max() noexcept { return 2147483647; }
+    static constexpr __device__ int min() noexcept { return -2147483648; }
+};
+#endif
+
 } // namespace std
 
 #else