From 00d24d254dcc1097555f3d4e893ba4cb17e49969 Mon Sep 17 00:00:00 2001 From: Aaron Enye Shi Date: Wed, 6 Mar 2019 18:33:43 +0000 Subject: [PATCH] Fix Agent_global variables failing hipTestDeviceSymbol Issue: Header uses std::vector agent_globals which is created by hip_module.cpp - Move iterator fails to copy Agent_global from library source into header version - Due to different versions of std::string name in struct Agent_global Fix: Change Agent_global to use char* name instead of std::string name --- include/hip/hcc_detail/hip_runtime_api.h | 4 ++-- src/hip_module.cpp | 7 +++++-- 2 files changed, 7 insertions(+), 4 deletions(-) diff --git a/include/hip/hcc_detail/hip_runtime_api.h b/include/hip/hcc_detail/hip_runtime_api.h index 72cfbf7d3b..f6ac507c6f 100644 --- a/include/hip/hcc_detail/hip_runtime_api.h +++ b/include/hip/hcc_detail/hip_runtime_api.h @@ -2506,7 +2506,7 @@ hipError_t hipModuleGetFunction(hipFunction_t* function, hipModule_t module, con hipError_t hipFuncGetAttributes(hipFuncAttributes* attr, const void* func); struct Agent_global { - std::string name; + char* name; hipDeviceptr_t address; uint32_t byte_cnt; }; @@ -2524,7 +2524,7 @@ template std::pair read_global_description( ForwardIterator f, ForwardIterator l, const char* name) { const auto it = std::find_if(f, l, [=](const Agent_global& x) { - return x.name == name; + return strcmp(x.name, name) == 0; }); return it == l ? diff --git a/src/hip_module.cpp b/src/hip_module.cpp index 178b0c9471..7a76c4642f 100644 --- a/src/hip_module.cpp +++ b/src/hip_module.cpp @@ -310,7 +310,7 @@ namespace hip_impl { namespace { inline void track(const Agent_global& x, hsa_agent_t agent) { - tprintf(DB_MEM, " add variable '%s' with ptr=%p size=%u to tracker\n", x.name.c_str(), + tprintf(DB_MEM, " add variable '%s' with ptr=%p size=%u to tracker\n", x.name, x.address, x.byte_cnt); int deviceIndex =0; @@ -341,7 +341,10 @@ inline hsa_status_t copy_agent_global_variables(hsa_executable_t, hsa_agent_t ag hsa_executable_symbol_get_info(x, HSA_EXECUTABLE_SYMBOL_INFO_TYPE, &t); if (t == HSA_SYMBOL_KIND_VARIABLE) { - static_cast(out)->push_back(Agent_global{name(x), address(x), size(x)}); + Agent_global tmp = {nullptr, address(x), size(x)}; + tmp.name = (char *) malloc(sizeof(name(x).c_str())); + strcpy(tmp.name, name(x).c_str()); + static_cast(out)->push_back(tmp); track(static_cast(out)->back(),agent); }