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

@@ -100,6 +100,7 @@ hipMemset2D
 hipMemsetAsync
 hipMemset2DAsync
 hipMemsetD8
+hipMemset3D
 hipModuleGetFunction
 hipModuleGetGlobal
 hipModuleLaunchKernel

api/hip/hip_hcc.map.in

@@ -101,6 +101,7 @@ global:
     hipMemsetAsync;
     hipMemset2DAsync;
     hipMemsetD8;
+    hipMemset3D;
     hipModuleGetFunction;
     hipModuleGetGlobal;
     hipModuleLaunchKernel;

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);