From cf4bdb8b55b5ead2fbdababa788c8ca444ada569 Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Fri, 5 Apr 2019 22:54:41 -0400 Subject: [PATCH] Fix regression on multi-gpu due to PR#997 [ROCm/hip commit: 271fdc4e4d88714e9b0eb6a02e19b15292a80e84] --- projects/hip/src/hip_module.cpp | 13 +++---------- 1 file changed, 3 insertions(+), 10 deletions(-) diff --git a/projects/hip/src/hip_module.cpp b/projects/hip/src/hip_module.cpp index c00004656a..05daca6055 100644 --- a/projects/hip/src/hip_module.cpp +++ b/projects/hip/src/hip_module.cpp @@ -361,13 +361,13 @@ inline hsa_status_t copy_agent_global_variables(hsa_executable_t, hsa_agent_t ag } hsa_executable_symbol_t find_kernel_by_name(hsa_executable_t executable, const char* kname, - hsa_agent_t agent) { + hsa_agent_t* agent = nullptr) { using namespace hip_impl; pair r{kname, {}}; hsa_executable_iterate_agent_symbols( - executable, agent, + executable, agent ? *agent : this_agent(), [](hsa_executable_t, hsa_agent_t, hsa_executable_symbol_t x, void* s) { auto p = static_cast*>(s); @@ -385,10 +385,6 @@ hsa_executable_symbol_t find_kernel_by_name(hsa_executable_t executable, const c return r.second; } -hsa_executable_symbol_t find_kernel_by_name(hsa_executable_t executable, const char* kname) { - return find_kernel_by_name(executable, kname, hip_impl::this_agent()); -} - string read_elf_file_as_string(const void* file) { // Precondition: file points to an ELF image that was BITWISE loaded @@ -457,10 +453,7 @@ hipError_t ihipModuleGetFunction(hipFunction_t* func, hipModule_t hmod, const ch if (!*func) return hipErrorInvalidValue; - if (!agent) - *agent = this_agent(); - - auto kernel = find_kernel_by_name(hmod->executable, name, *agent); + auto kernel = find_kernel_by_name(hmod->executable, name, agent); if (kernel.handle == 0u) return hipErrorNotFound;