From 4fcfb64a8848314a0dc2d5d31e19b3eadbc022c7 Mon Sep 17 00:00:00 2001
From: Alex Voicu <alexandru.voicu@amd.com>
Date: Mon, 25 Jun 2018 19:16:27 +0100
Subject: [PATCH] Removes use of unimplemented OCML functionality.

[ROCm/clr commit: 9452298fbe29bf53cad18aa1d368d3345d5eaa9f]
---
 .../include/hip/hcc_detail/math_functions.h   | 18 ++++++-------
 .../hipamd/include/hip/hcc_detail/math_fwd.h  | 26 -------------------
 2 files changed, 9 insertions(+), 35 deletions(-)

diff --git a/projects/clr/hipamd/include/hip/hcc_detail/math_functions.h b/projects/clr/hipamd/include/hip/hcc_detail/math_functions.h
index a9288dc204..8f7c885fc7 100644
--- a/projects/clr/hipamd/include/hip/hcc_detail/math_functions.h
+++ b/projects/clr/hipamd/include/hip/hcc_detail/math_functions.h
@@ -508,19 +508,19 @@ inline
 float __fadd_rz(float x, float y) { return __ocml_add_rtz_f32(x, y); }
 __device__
 inline
-float __fdiv_rd(float x, float y) { return __ocml_div_rtp_f32(x, y); }
+float __fdiv_rd(float x, float y) { return x / y; }
 __device__
 inline
-float __fdiv_rn(float x, float y) { return __ocml_div_rte_f32(x, y); }
+float __fdiv_rn(float x, float y) { return x / y; }
 __device__
 inline
-float __fdiv_ru(float x, float y) { return __ocml_div_rtn_f32(x, y); }
+float __fdiv_ru(float x, float y) { return x / y; }
 __device__
 inline
-float __fdiv_rz(float x, float y) { return __ocml_div_rtz_f32(x, y); }
+float __fdiv_rz(float x, float y) { return x / y; }
 __device__
 inline
-float __fdividef(float x, float y) { return __ocml_div_rte_f32(x, y); }
+float __fdividef(float x, float y) { return x / y; }
 __device__
 inline
 float __fmaf_rd(float x, float y, float z)
@@ -1028,16 +1028,16 @@ inline
 double __dadd_rz(double x, double y) { return __ocml_add_rtz_f64(x, y); }
 __device__
 inline
-double __ddiv_rd(double x, double y) { return __ocml_div_rtp_f64(x, y); }
+double __ddiv_rd(double x, double y) { return x / y; }
 __device__
 inline
-double __ddiv_rn(double x, double y) { return __ocml_div_rte_f64(x, y); }
+double __ddiv_rn(double x, double y) { return x / y; }
 __device__
 inline
-double __ddiv_ru(double x, double y) { return __ocml_div_rtn_f64(x, y); }
+double __ddiv_ru(double x, double y) { return x / y; }
 __device__
 inline
-double __ddiv_rz(double x, double y) { return __ocml_div_rtz_f64(x, y); }
+double __ddiv_rz(double x, double y) { return x / y; }
 __device__
 inline
 double __dmul_rd(double x, double y) { return __ocml_mul_rtp_f64(x, y); }
diff --git a/projects/clr/hipamd/include/hip/hcc_detail/math_fwd.h b/projects/clr/hipamd/include/hip/hcc_detail/math_fwd.h
index ac5d6e7dc6..102714e117 100644
--- a/projects/clr/hipamd/include/hip/hcc_detail/math_fwd.h
+++ b/projects/clr/hipamd/include/hip/hcc_detail/math_fwd.h
@@ -69,8 +69,6 @@ float __ocml_cosh_f32(float);
 __device__
 float __ocml_cospi_f32(float);
 __device__
-float __ocml_div_rtz_f32(float, float);
-__device__
 float __ocml_i0_f32(float);
 __device__
 float __ocml_i1_f32(float);
@@ -290,18 +288,6 @@ __attribute__((const))
 float __ocml_mul_rtz_f32(float, float);
 __device__
 __attribute__((const))
-float __ocml_div_rte_f32(float, float);
-__device__
-__attribute__((const))
-float __ocml_div_rtn_f32(float, float);
-__device__
-__attribute__((const))
-float __ocml_div_rtp_f32(float, float);
-__device__
-__attribute__((const))
-float __ocml_div_rtz_f32(float, float);
-__device__
-__attribute__((const))
 float __ocml_sqrt_rte_f32(float);
 __device__
 __attribute__((const))
@@ -598,18 +584,6 @@ __attribute__((const))
 double __ocml_mul_rtz_f64(double, double);
 __device__
 __attribute__((const))
-double __ocml_div_rte_f64(double, double);
-__device__
-__attribute__((const))
-double __ocml_div_rtn_f64(double, double);
-__device__
-__attribute__((const))
-double __ocml_div_rtp_f64(double, double);
-__device__
-__attribute__((const))
-double __ocml_div_rtz_f64(double, double);
-__device__
-__attribute__((const))
 double __ocml_sqrt_rte_f64(double);
 __device__
 __attribute__((const))