From 86dd262e9be94c5a7afcb91c8ce2c0969d3dfbc8 Mon Sep 17 00:00:00 2001 From: Laurent Morichetti Date: Tue, 29 Oct 2019 08:23:57 -0700 Subject: [PATCH] Fix a code object memory corruption The lifetime of the buffer given to hsa_code_object_reader_create_from_memory must exceed that of the code object reader. We need to create a copy of the code object binary memory (file) that is kept allocated until the code object reader is destroyed. [ROCm/hip commit: 7473140a76472020b77c5d6414a9f12acb6702bd] --- projects/hip/src/program_state.inl | 12 +++++++++--- 1 file changed, 9 insertions(+), 3 deletions(-) diff --git a/projects/hip/src/program_state.inl b/projects/hip/src/program_state.inl index 9feabbc2f7..3dc0a9e128 100644 --- a/projects/hip/src/program_state.inl +++ b/projects/hip/src/program_state.inl @@ -200,7 +200,7 @@ public: std::function>; std::pair< std::mutex, - std::vector> code_readers; + std::vector, RAII_code_reader>>> code_readers; program_state_impl() { // Create placeholder for each agent for the per-agent members. @@ -390,8 +390,14 @@ public: }; RAII_code_reader tmp{new hsa_code_object_reader_t, cor_deleter}; + // The lifetime of the buffer must exceed that of the associated + // code object reader. Create a copy of the file, which will be + // released at the same time the code object reader is destroyed. + std::unique_ptr buffer(new char[file.size()]); + memcpy(buffer.get(), file.data(), file.size()); + hsa_code_object_reader_create_from_memory( - file.data(), file.size(), tmp.get()); + buffer.get(), file.size(), tmp.get()); hsa_executable_load_agent_code_object( executable, agent, *tmp, nullptr, nullptr); @@ -399,7 +405,7 @@ public: hsa_executable_freeze(executable, nullptr); std::lock_guard lck{code_readers.first}; - code_readers.second.push_back(move(tmp)); + code_readers.second.emplace_back(move(buffer), move(tmp)); }