From c2ffca9c1dae7ce6a7a03247c40ead1f1dc59b09 Mon Sep 17 00:00:00 2001
From: foreman <dl.swbuild@amd.com>
Date: Mon, 12 Aug 2019 21:22:47 -0400
Subject: [PATCH] P4 to Git Change 1982056 by kjayapra@3_HIPWS_TXT_ROCM on
 2019/08/12 21:16:51

	SWDEV-199635 - Setting VarInfoCallBack before all program builds.

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_internal.hpp#36 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_module.cpp#33 edit
... //depot/stg/opencl/drivers/opencl/runtime/device/devprogram.cpp#58 edit
... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palprogram.cpp#97 edit
... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocprogram.cpp#106 edit
---
 api/hip/hip_internal.hpp | 2 ++
 api/hip/hip_module.cpp   | 2 ++
 2 files changed, 4 insertions(+)

diff --git a/api/hip/hip_internal.hpp b/api/hip/hip_internal.hpp
index 1f7e31c9aa..0df562065c 100644
--- a/api/hip/hip_internal.hpp
+++ b/api/hip/hip_internal.hpp
@@ -175,6 +175,8 @@ extern hipError_t ihipDeviceGetCount(int* count);
 extern int ihipGetDevice();
 extern hipError_t ihipMalloc(void** ptr, size_t sizeBytes, unsigned int flags);
 extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset);
+extern bool CL_CALLBACK getSvarInfo(cl_program program, std::string var_name, void** var_addr,
+                                    size_t* var_size);
 
 #define HIP_RETURN(ret)          \
         hip::g_lastError = ret;  \
diff --git a/api/hip/hip_module.cpp b/api/hip/hip_module.cpp
index 77a8a34659..a66c2da9e9 100644
--- a/api/hip/hip_module.cpp
+++ b/api/hip/hip_module.cpp
@@ -143,6 +143,8 @@ hipError_t ihipModuleLoadData(hipModule_t *module, const void *image)
     return hipErrorOutOfMemory;
   }
 
+  program->setVarInfoCallBack(&getSvarInfo);
+
   if (CL_SUCCESS != program->addDeviceProgram(*hip::getCurrentContext()->devices()[0], image, ElfSize(image)) ||
     CL_SUCCESS != program->build(hip::getCurrentContext()->devices(), nullptr, nullptr, nullptr)) {
     return hipErrorUnknown;