AI/rocm-systems
2
0
Çatalla 0
Kod Konular Değişiklik İstekleri İşlemler Paketler Projeler Sürüm Viki Aktivite

Remove load_legacy parameter + change prefix for some loaded code object queries back to AMD

Kaynağa Gözat 
Change-Id: I74e905abd77dab3a7a00b5ced94cd9b5130365c5


[ROCm/ROCR-Runtime commit: 4b86843409]
Bu işleme şunda yer alıyor:
Konstantin Zhuravlyov
2016-11-20 13:46:17 -05:00
ebeveyn 2e8c176e11
işleme 05a7b5a9e4
4 değiştirilmiş dosya ile 34 ekleme ve 36 silme
Yama Dosyasını İndir Diff Dosyasını İndir Tüm dosyaları genişlet Tüm dosyaları daralt
projects/rocr-runtime/runtime/hsa-runtime/core/inc/amd_hsa_loader.hpp
+6 -8
Dosyayı Görüntüle
@@ -74,9 +74,9 @@ typedef hsa_executable_symbol_t hsa_symbol_t;
typedef hsa_executable_symbol_info_t hsa_symbol_info_t;
/// @brief Loaded code object attributes.
enum hsa_loaded_code_object_info_t {
HSA_LOADED_CODE_OBJECT_INFO_ELF_IMAGE = 0,
HSA_LOADED_CODE_OBJECT_INFO_ELF_IMAGE_SIZE = 1
enum amd_loaded_code_object_info_t {
AMD_LOADED_CODE_OBJECT_INFO_ELF_IMAGE = 0,
AMD_LOADED_CODE_OBJECT_INFO_ELF_IMAGE_SIZE = 1
};
/// @brief Loaded segment handle.
@@ -200,7 +200,7 @@ public:
virtual ~LoadedCodeObject() {}
virtual bool GetInfo(hsa_loaded_code_object_info_t attribute, void *value) = 0;
virtual bool GetInfo(amd_loaded_code_object_info_t attribute, void *value) = 0;
virtual hsa_status_t IterateLoadedSegments(
hsa_status_t (*callback)(
@@ -282,16 +282,14 @@ public:
hsa_agent_t agent,
hsa_code_object_t code_object,
const char *options,
hsa_loaded_code_object_t *loaded_code_object = nullptr,
bool load_legacy = true) = 0;
hsa_loaded_code_object_t *loaded_code_object = nullptr) = 0;
virtual hsa_status_t LoadCodeObject(
hsa_agent_t agent,
hsa_code_object_t code_object,
size_t code_object_size,
const char *options,
hsa_loaded_code_object_t *loaded_code_object = nullptr,
bool load_legacy = true) = 0;
hsa_loaded_code_object_t *loaded_code_object = nullptr) = 0;
virtual hsa_status_t Freeze(const char *options) = 0;
projects/rocr-runtime/runtime/hsa-runtime/core/runtime/hsa.cpp
+2 -2
Dosyayı Görüntüle
@@ -1926,7 +1926,7 @@ hsa_status_t hsa_executable_load_program_code_object(
hsa_code_object_t code_object =
{reinterpret_cast<uint64_t>(wrapper->code_object_memory)};
return exec->LoadCodeObject(
{0}, code_object, options, loaded_code_object, false);
{0}, code_object, options, loaded_code_object);
}
hsa_status_t hsa_executable_load_agent_code_object(
@@ -1951,7 +1951,7 @@ hsa_status_t hsa_executable_load_agent_code_object(
hsa_code_object_t code_object =
{reinterpret_cast<uint64_t>(wrapper->code_object_memory)};
return exec->LoadCodeObject(
agent, code_object, options, loaded_code_object, false);
agent, code_object, options, loaded_code_object);
}
hsa_status_t hsa_executable_freeze(
projects/rocr-runtime/runtime/hsa-runtime/loader/executable.cpp
+20 -17
Dosyayı Görüntüle
@@ -529,15 +529,15 @@ bool VariableSymbol::GetInfo(hsa_symbol_info32_t symbol_info, void *value) {
return true;
}
bool LoadedCodeObjectImpl::GetInfo(hsa_loaded_code_object_info_t attribute, void *value)
bool LoadedCodeObjectImpl::GetInfo(amd_loaded_code_object_info_t attribute, void *value)
{
assert(value);
switch (attribute) {
case HSA_LOADED_CODE_OBJECT_INFO_ELF_IMAGE:
case AMD_LOADED_CODE_OBJECT_INFO_ELF_IMAGE:
((hsa_code_object_t*)value)->handle = reinterpret_cast<uint64_t>(elf_data);
break;
case HSA_LOADED_CODE_OBJECT_INFO_ELF_IMAGE_SIZE:
case AMD_LOADED_CODE_OBJECT_INFO_ELF_IMAGE_SIZE:
*((size_t*)value) = elf_size;
break;
default: {
@@ -988,10 +988,9 @@ hsa_status_t ExecutableImpl::LoadCodeObject(
hsa_agent_t agent,
hsa_code_object_t code_object,
const char *options,
hsa_loaded_code_object_t *loaded_code_object,
bool load_legacy)
hsa_loaded_code_object_t *loaded_code_object)
{
return LoadCodeObject(agent, code_object, 0, options, loaded_code_object, load_legacy);
return LoadCodeObject(agent, code_object, 0, options, loaded_code_object);
}
hsa_status_t ExecutableImpl::LoadCodeObject(
@@ -999,8 +998,7 @@ hsa_status_t ExecutableImpl::LoadCodeObject(
hsa_code_object_t code_object,
size_t code_object_size,
const char *options,
hsa_loaded_code_object_t *loaded_code_object,
bool load_legacy)
hsa_loaded_code_object_t *loaded_code_object)
{
WriterLockGuard<ReaderWriterLock> writer_lock(rw_lock_);
if (HSA_EXECUTABLE_STATE_FROZEN == state_) {
@@ -1018,7 +1016,6 @@ hsa_status_t ExecutableImpl::LoadCodeObject(
}
code.reset(new code::AmdHsaCode());
load_legacy_ = load_legacy;
if (!code->InitAsHandle(code_object)) {
return HSA_STATUS_ERROR_INVALID_CODE_OBJECT;
@@ -1079,7 +1076,7 @@ hsa_status_t ExecutableImpl::LoadCodeObject(
if (status != HSA_STATUS_SUCCESS) return status;
for (size_t i = 0; i < code->SymbolCount(); ++i) {
status = LoadSymbol(agent, code->GetSymbol(i));
status = LoadSymbol(agent, code->GetSymbol(i), majorVersion);
if (status != HSA_STATUS_SUCCESS) { return status; }
}
@@ -1185,20 +1182,24 @@ hsa_status_t ExecutableImpl::LoadSegmentV2(const code::Segment *data_segment,
return HSA_STATUS_SUCCESS;
}
hsa_status_t ExecutableImpl::LoadSymbol(hsa_agent_t agent, code::Symbol* sym)
hsa_status_t ExecutableImpl::LoadSymbol(hsa_agent_t agent,
code::Symbol* sym,
uint32_t majorVersion)
{
if (sym->IsDeclaration()) {
return LoadDeclarationSymbol(agent, sym);
return LoadDeclarationSymbol(agent, sym, majorVersion);
} else {
return LoadDefinitionSymbol(agent, sym);
return LoadDefinitionSymbol(agent, sym, majorVersion);
}
}
hsa_status_t ExecutableImpl::LoadDefinitionSymbol(hsa_agent_t agent, code::Symbol* sym)
hsa_status_t ExecutableImpl::LoadDefinitionSymbol(hsa_agent_t agent,
code::Symbol* sym,
uint32_t majorVersion)
{
bool isAgent = sym->IsAgent();
if (!load_legacy_) {
isAgent = agent.handle == 0 ? false : true;
if (majorVersion >= 2) {
isAgent = agent.handle != 0;
}
if (isAgent) {
auto agent_symbol = agent_symbols_.find(std::make_pair(sym->Name(), agent));
@@ -1289,7 +1290,9 @@ hsa_status_t ExecutableImpl::LoadDefinitionSymbol(hsa_agent_t agent, code::Symbo
return HSA_STATUS_SUCCESS;
}
hsa_status_t ExecutableImpl::LoadDeclarationSymbol(hsa_agent_t agent, code::Symbol* sym)
hsa_status_t ExecutableImpl::LoadDeclarationSymbol(hsa_agent_t agent,
code::Symbol* sym,
uint32_t majorVersion)
{
auto program_symbol = program_symbols_.find(sym->Name());
if (program_symbol == program_symbols_.end()) {
projects/rocr-runtime/runtime/hsa-runtime/loader/executable.hpp
+6 -9
Dosyayı Görüntüle
@@ -257,7 +257,7 @@ public:
size_t ElfSize() const { return elf_size; }
std::vector<Segment*>& LoadedSegments() { return loaded_segments; }
bool GetInfo(hsa_loaded_code_object_info_t attribute, void *value) override;
bool GetInfo(amd_loaded_code_object_info_t attribute, void *value) override;
hsa_status_t IterateLoadedSegments(
hsa_status_t (*callback)(
@@ -376,16 +376,14 @@ public:
hsa_agent_t agent,
hsa_code_object_t code_object,
const char *options,
hsa_loaded_code_object_t *loaded_code_object,
bool load_legacy = true) override;
hsa_loaded_code_object_t *loaded_code_object) override;
hsa_status_t LoadCodeObject(
hsa_agent_t agent,
hsa_code_object_t code_object,
size_t code_object_size,
const char *options,
hsa_loaded_code_object_t *loaded_code_object,
bool load_legacy = true) override;
hsa_loaded_code_object_t *loaded_code_object) override;
hsa_status_t Freeze(const char *options) override;
@@ -452,7 +450,6 @@ private:
ExecutableImpl& operator=(const ExecutableImpl &e);
std::unique_ptr<amd::hsa::code::AmdHsaCode> code;
bool load_legacy_;
Symbol* GetSymbolInternal(
const char *symbol_name,
@@ -466,9 +463,9 @@ private:
hsa_status_t LoadSegmentV2(const code::Segment *data_segment,
loader::Segment *load_segment);
hsa_status_t LoadSymbol(hsa_agent_t agent, amd::hsa::code::Symbol* sym);
hsa_status_t LoadDefinitionSymbol(hsa_agent_t agent, amd::hsa::code::Symbol* sym);
hsa_status_t LoadDeclarationSymbol(hsa_agent_t agent, amd::hsa::code::Symbol* sym);
hsa_status_t LoadSymbol(hsa_agent_t agent, amd::hsa::code::Symbol* sym, uint32_t majorVersion);
hsa_status_t LoadDefinitionSymbol(hsa_agent_t agent, amd::hsa::code::Symbol* sym, uint32_t majorVersion);
hsa_status_t LoadDeclarationSymbol(hsa_agent_t agent, amd::hsa::code::Symbol* sym, uint32_t majorVersion);
hsa_status_t ApplyRelocations(hsa_agent_t agent, amd::hsa::code::AmdHsaCode *c);
hsa_status_t ApplyStaticRelocationSection(hsa_agent_t agent, amd::hsa::code::RelocationSection* sec);