P4 to Git Change 1541197 by skudchad@skudchad_test2_win_opencl on 2018/04/13 17:49:07

SWDEV-145570 - [HIP] -  Add missing HIP APIs and  fill sym table references. The HIP tests can now build if we do a symlink to libhip_hcc.so that we build with VDI in the install folder in github HIP.

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

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_device.cpp#10 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_hcc.def.in#3 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_hcc.map.in#3 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_internal.hpp#7 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_platform.cpp#5 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_profile.cpp#1 add
... //depot/stg/opencl/drivers/opencl/api/hip/hip_texture.cpp#3 edit

hipamd/api/hip/hip_device.cpp

@@ -191,3 +191,18 @@ hipError_t hipGetDeviceProperties ( hipDeviceProp_t* props, hipDevice_t device )
   return hipSuccess;
 }
 
+hipError_t hipHccGetAccelerator(int deviceId, hc::accelerator* acc) {
+  HIP_INIT_API(deviceId, acc);
+
+  assert(0 && "Unimplemented");
+
+  return hipErrorUnknown;
+}
+
+hipError_t hipHccGetAcceleratorView(hipStream_t stream, hc::accelerator_view** av) {
+  HIP_INIT_API(stream, av);
+
+  assert(0 && "Unimplemented");
+
+  return hipErrorUnknown;
+}

hipamd/api/hip/hip_hcc.def.in

@@ -89,6 +89,7 @@ hipMemcpyHtoDAsync
 hipMemcpyPeer
 hipMemcpyPeerAsync
 hipMemcpyToArray
+hipMemcpyFromArray
 hipMemcpyToSymbol
 hipMemcpyToSymbolAsync
 hipMemGetAddressRange
@@ -127,3 +128,30 @@ __hipUnregisterFatBinary
 hipConfigureCall
 hipSetupArgument
 hipLaunchByPtr
+hipCreateTextureObject
+hipDestroyTextureObject
+hipGetTextureObjectResourceDesc
+hipGetTextureObjectResourceViewDesc
+hipGetTextureObjectTextureDesc
+hipBindTexture
+hipBindTexture2D
+hipBindTextureToArray
+hipBindTextureToMipmappedArray
+hipUnbindTexture
+hipGetChannelDesc
+hipGetTextureAlignmentOffset
+hipGetTextureReference
+hipTexRefSetFormat
+hipTexRefSetFlags
+hipTexRefSetFilterMode
+hipTexRefSetAddressMode
+hipTexRefSetArray
+hipTexRefSetAddress
+hipTexRefSetAddress2D
+hipCreateChannelDesc
+hipProfilerStart
+hipProfilerStop
+hipHccGetAccelerator
+hipHccGetAcceleratorView
+hipCreateSurfaceObject
+hipDestroySurfaceObject

hipamd/api/hip/hip_hcc.map.in

@@ -90,6 +90,7 @@ global:
     hipMemcpyPeer;
     hipMemcpyPeerAsync;
     hipMemcpyToArray;
+    hipMemcpyFromArray;
     hipMemcpyToSymbol;
     hipMemcpyToSymbolAsync;
     hipMemGetAddressRange;
@@ -128,6 +129,37 @@ global:
     hipConfigureCall;
     hipSetupArgument;
     hipLaunchByPtr;
+    hipProfilerStart;
+    hipProfilerStop;
+    extern "C++" {
+    hip_impl::hipLaunchKernelGGLImpl*;
+    hipCreateTextureObject*;
+    hipDestroyTextureObject*;
+    hipGetTextureObjectResourceDesc;
+    hipGetTextureObjectResourceViewDesc;
+    hipGetTextureObjectTextureDesc;
+    hipBindTexture;
+    hipBindTexture2D;
+    hipBindTextureToArray;
+    hipBindTextureToMipmappedArray;
+    hipUnbindTexture;
+    hipGetChannelDesc;
+    hipGetTextureAlignmentOffset;
+    hipGetTextureReference;
+    hipTexRefSetFormat;
+    hipTexRefSetFlags;
+    hipTexRefSetFilterMode;
+    hipTexRefSetAddressMode;
+    hipTexRefSetArray;
+    hipTexRefSetAddress;
+    hipTexRefSetAddress2D;
+    hipCreateChannelDesc*;
+    ihipBindTextureToArrayImpl*;
+    hipHccGetAccelerator*;
+    hipHccGetAcceleratorView*;
+    hipCreateSurfaceObject*;
+    hipDestroySurfaceObject*;
+    };
 local:
     *;
 };

hipamd/api/hip/hip_internal.hpp

@@ -40,6 +40,11 @@ THE SOFTWARE.
     return hipErrorOutOfMemory;                              \
   }
 
+namespace hc {
+class accelerator;
+class accelerator_view;
+};
+
 extern std::once_flag g_ihipInitialized;
 extern thread_local amd::Context* g_context;
 extern std::vector<amd::Context*> g_devices;

hipamd/api/hip/hip_platform.cpp

@@ -260,3 +260,20 @@ extern "C" hipError_t hipLaunchByPtr(const void *hostFunction)
     g_blockDim.x, g_blockDim.y, g_blockDim.z,
     g_sharedMem, g_stream, nullptr, extra);
 }
+
+#if defined(ATI_OS_LINUX)
+
+namespace hip_impl {
+
+void hipLaunchKernelGGLImpl(
+  uintptr_t function_address,
+  const dim3& numBlocks,
+  const dim3& dimBlocks,
+  uint32_t sharedMemBytes,
+  hipStream_t stream,
+  void** kernarg) {
+}
+
+}
+
+#endif // defined(ATI_OS_LINUX)

hipamd/api/hip/hip_profile.cpp

@@ -0,0 +1,42 @@
+/*
+Copyright (c) 2015 - present Advanced Micro Devices, Inc. All rights reserved.
+
+Permission is hereby granted, free of charge, to any person obtaining a copy
+of this software and associated documentation files (the "Software"), to deal
+in the Software without restriction, including without limitation the rights
+to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+copies of the Software, and to permit persons to whom the Software is
+furnished to do so, subject to the following conditions:
+
+The above copyright notice and this permission notice shall be included in
+all copies or substantial portions of the Software.
+
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL THE
+AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+THE SOFTWARE.
+*/
+
+#include <hip/hip_runtime.h>
+
+#include "hip_internal.hpp"
+
+hipError_t hipProfilerStart() {
+  HIP_INIT_API();
+
+  assert(0 && "Unimplemented");
+
+  return hipErrorUnknown;
+}
+
+
+hipError_t hipProfilerStop() {
+  HIP_INIT_API();
+
+  assert(0 && "Unimplemented");
+
+  return hipErrorUnknown;
+}

hipamd/api/hip/hip_texture.cpp

@@ -98,6 +98,15 @@ hipError_t hipBindTextureToArray(textureReference* tex, hipArray_const_t array,
   return hipErrorUnknown;
 }
 
+hipError_t ihipBindTextureToArrayImpl(int dim, enum hipTextureReadMode readMode,
+                                      hipArray_const_t array,
+                                      const struct hipChannelFormatDesc& desc,
+                                      textureReference* tex) {
+  assert(0 && "Unimplemented");
+
+  return hipErrorUnknown;
+}
+
 hipError_t hipBindTextureToMipmappedArray(textureReference* tex,
                                           hipMipmappedArray_const_t mipmappedArray,
                                           const hipChannelFormatDesc* desc) {