From 638f62b0c49839cb19563c47f942a9c614abd70a Mon Sep 17 00:00:00 2001
From: Siu Chi Chan <siuchi.chan@amd.com>
Date: Tue, 24 Oct 2017 00:55:41 -0400
Subject: [PATCH] printf support for module API

[ROCm/hip commit: 1ddee10c2f9a6e8a5f9634c78c9c5c8b005093c4]
---
 projects/hip/src/hip_module.cpp | 5 +++++
 1 file changed, 5 insertions(+)

diff --git a/projects/hip/src/hip_module.cpp b/projects/hip/src/hip_module.cpp
index 74b0092b84..85c88a1826 100644
--- a/projects/hip/src/hip_module.cpp
+++ b/projects/hip/src/hip_module.cpp
@@ -259,6 +259,11 @@ hipError_t hipModuleLoad(hipModule_t *module, const char *fname){
             status = hsa_executable_create(HSA_PROFILE_FULL, HSA_EXECUTABLE_STATE_UNFROZEN, NULL, &(*module)->executable);
             CHECKLOG_HSA(status, hipErrorNotInitialized);
 
+            status = hsa_executable_agent_global_variable_define((*module)->executable, agent,
+                                                                 "_ZN2hc13printf_bufferE",
+                                                                 Kalmar::getContext()->getPrintfBufferPointerVA());
+            CHECKLOG_HSA(status, hipErrorNotInitialized);
+
             status = hsa_executable_load_code_object((*module)->executable, agent, (*module)->object, NULL);
             CHECKLOG_HSA(status, hipErrorNotInitialized);