From 7330137b0b1042e7e67f1f42f51225b6fbf2c0a5 Mon Sep 17 00:00:00 2001
From: foreman <dl.constructicon@amd.com>
Date: Wed, 13 Jun 2018 13:36:52 -0400
Subject: [PATCH] P4 to Git Change 1567831 by
 skudchad@skudchad_test2_win_opencl on 2018/06/13 13:20:10

	SWDEV-145570 - [HIP] - Implement hipMemset3D

	ReviewBoardURL = http://ocltc.amd.com/reviews/r/15164/diff/

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_hcc.def.in#5 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_hcc.map.in#5 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_memory.cpp#33 edit
---
 api/hip/hip_hcc.def.in |  1 +
 api/hip/hip_hcc.map.in |  1 +
 api/hip/hip_memory.cpp | 12 ++++++++++++
 3 files changed, 14 insertions(+)

diff --git a/api/hip/hip_hcc.def.in b/api/hip/hip_hcc.def.in
index d2916f3885..bb9c179d3b 100644
--- a/api/hip/hip_hcc.def.in
+++ b/api/hip/hip_hcc.def.in
@@ -100,6 +100,7 @@ hipMemset2D
 hipMemsetAsync
 hipMemset2DAsync
 hipMemsetD8
+hipMemset3D
 hipModuleGetFunction
 hipModuleGetGlobal
 hipModuleLaunchKernel
diff --git a/api/hip/hip_hcc.map.in b/api/hip/hip_hcc.map.in
index 761641425e..452f260534 100644
--- a/api/hip/hip_hcc.map.in
+++ b/api/hip/hip_hcc.map.in
@@ -101,6 +101,7 @@ global:
     hipMemsetAsync;
     hipMemset2DAsync;
     hipMemsetD8;
+    hipMemset3D;
     hipModuleGetFunction;
     hipModuleGetGlobal;
     hipModuleLaunchKernel;
diff --git a/api/hip/hip_memory.cpp b/api/hip/hip_memory.cpp
index cd4c60c319..376b2b155f 100644
--- a/api/hip/hip_memory.cpp
+++ b/api/hip/hip_memory.cpp
@@ -363,6 +363,18 @@ hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) {
   return status;
 }
 
+hipError_t hipMemset3D(hipPitchedPtr pitchedDevPtr, int value, hipExtent extent) {
+  HIP_INIT_API(pitchedDevPtr, value, &extent);
+
+  void *dst = &pitchedDevPtr.ptr;
+  size_t sizeBytes = pitchedDevPtr.pitch * extent.height * extent.depth;
+
+  hip::syncStreams();
+  amd::HostQueue* queue = hip::getNullStream();
+
+  return ihipMemset(&dst, value, sizeBytes, *queue);
+}
+
 hipError_t hipArrayCreate(hipArray** array, const HIP_ARRAY_DESCRIPTOR* pAllocateArray) {
   HIP_INIT_API(array, pAllocateArray);