diff --git a/llvm/lib/ObjectYAML/ELFYAML.cpp b/llvm/lib/ObjectYAML/ELFYAML.cpp index 70dc2d10735d8..887c27a0b8fe6 100644 --- a/llvm/lib/ObjectYAML/ELFYAML.cpp +++ b/llvm/lib/ObjectYAML/ELFYAML.cpp @@ -155,6 +155,9 @@ void ScalarEnumerationTraits::enumeration( ECase(NT_FREEBSD_PROCSTAT_PSSTRINGS); ECase(NT_FREEBSD_PROCSTAT_AUXV); // AMD specific notes. (Code Object V2) + ECase(NT_AMD_HSA_CODE_OBJECT_VERSION); + ECase(NT_AMD_HSA_HSAIL); + ECase(NT_AMD_HSA_ISA_VERSION); ECase(NT_AMD_HSA_METADATA); ECase(NT_AMD_HSA_ISA_NAME); ECase(NT_AMD_PAL_METADATA); diff --git a/llvm/test/tools/llvm-readobj/ELF/note-amd-invalid-v2.test b/llvm/test/tools/llvm-readobj/ELF/note-amd-invalid-v2.test new file mode 100644 index 0000000000000..778724b8ab6ce --- /dev/null +++ b/llvm/test/tools/llvm-readobj/ELF/note-amd-invalid-v2.test @@ -0,0 +1,233 @@ +## This test is checking the handling of invalid/malformed note entries for +## AMDGPU code object v2. + +# RUN: yaml2obj %s -o %t.o +# RUN: llvm-readobj --notes %t.o | FileCheck %s --match-full-lines --check-prefix=LLVM +# RUN: llvm-readelf --notes %t.o | FileCheck %s --match-full-lines --check-prefix=GNU + +# LLVM: Notes [ +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_code_object_version_0 +# LLVM-NEXT: Offset: 0x40 +# LLVM-NEXT: Size: 0x14 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x4 +# LLVM-NEXT: Type: NT_AMD_HSA_CODE_OBJECT_VERSION (AMD HSA Code Object Version) +# LLVM-NEXT: AMD HSA Code Object Version: Invalid AMD HSA Code Object Version +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_code_object_version_1 +# LLVM-NEXT: Offset: 0x54 +# LLVM-NEXT: Size: 0x1C +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0xC +# LLVM-NEXT: Type: NT_AMD_HSA_CODE_OBJECT_VERSION (AMD HSA Code Object Version) +# LLVM-NEXT: AMD HSA Code Object Version: Invalid AMD HSA Code Object Version +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_hsail_0 +# LLVM-NEXT: Offset: 0x70 +# LLVM-NEXT: Size: 0x1C +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0xA +# LLVM-NEXT: Type: NT_AMD_HSA_HSAIL (AMD HSA HSAIL Properties) +# LLVM-NEXT: AMD HSA HSAIL Properties: Invalid AMD HSA HSAIL Properties +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_hsail_1 +# LLVM-NEXT: Offset: 0x8C +# LLVM-NEXT: Size: 0x24 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x14 +# LLVM-NEXT: Type: NT_AMD_HSA_HSAIL (AMD HSA HSAIL Properties) +# LLVM-NEXT: AMD HSA HSAIL Properties: Invalid AMD HSA HSAIL Properties +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_isa_version_0 +# LLVM-NEXT: Offset: 0xB0 +# LLVM-NEXT: Size: 0x18 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x8 +# LLVM-NEXT: Type: NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# LLVM-NEXT: AMD HSA ISA Version: Invalid AMD HSA ISA Version +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_isa_version_1 +# LLVM-NEXT: Offset: 0xC8 +# LLVM-NEXT: Size: 0x28 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x17 +# LLVM-NEXT: Type: NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# LLVM-NEXT: AMD HSA ISA Version: Invalid AMD HSA ISA Version +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_isa_version_2 +# LLVM-NEXT: Offset: 0xF0 +# LLVM-NEXT: Size: 0x28 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x17 +# LLVM-NEXT: Type: NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# LLVM-NEXT: AMD HSA ISA Version: Invalid AMD HSA ISA Version +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_isa_version_3 +# LLVM-NEXT: Offset: 0x118 +# LLVM-NEXT: Size: 0x28 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x17 +# LLVM-NEXT: Type: NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# LLVM-NEXT: AMD HSA ISA Version: Invalid AMD HSA ISA Version +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_pal_metadata +# LLVM-NEXT: Offset: 0x140 +# LLVM-NEXT: Size: 0x14 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x4 +# LLVM-NEXT: Type: NT_AMD_PAL_METADATA (AMD PAL Metadata) +# LLVM-NEXT: AMD PAL Metadata: Invalid AMD PAL Metadata +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: ] + +# GNU: Displaying notes found in: .note.nt_amd_hsa_code_object_version_0 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000004 NT_AMD_HSA_CODE_OBJECT_VERSION (AMD HSA Code Object Version) +# GNU-NEXT: AMD HSA Code Object Version: +# GNU-NEXT: Invalid AMD HSA Code Object Version +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_code_object_version_1 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x0000000c NT_AMD_HSA_CODE_OBJECT_VERSION (AMD HSA Code Object Version) +# GNU-NEXT: AMD HSA Code Object Version: +# GNU-NEXT: Invalid AMD HSA Code Object Version +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_hsail_0 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x0000000a NT_AMD_HSA_HSAIL (AMD HSA HSAIL Properties) +# GNU-NEXT: AMD HSA HSAIL Properties: +# GNU-NEXT: Invalid AMD HSA HSAIL Properties +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_hsail_1 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000014 NT_AMD_HSA_HSAIL (AMD HSA HSAIL Properties) +# GNU-NEXT: AMD HSA HSAIL Properties: +# GNU-NEXT: Invalid AMD HSA HSAIL Properties +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_isa_version_0 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000008 NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# GNU-NEXT: AMD HSA ISA Version: +# GNU-NEXT: Invalid AMD HSA ISA Version +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_isa_version_1 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000017 NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# GNU-NEXT: AMD HSA ISA Version: +# GNU-NEXT: Invalid AMD HSA ISA Version +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_isa_version_2 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000017 NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# GNU-NEXT: AMD HSA ISA Version: +# GNU-NEXT: Invalid AMD HSA ISA Version +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_isa_version_3 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000017 NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# GNU-NEXT: AMD HSA ISA Version: +# GNU-NEXT: Invalid AMD HSA ISA Version +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_pal_metadata +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000004 NT_AMD_PAL_METADATA (AMD PAL Metadata) +# GNU-NEXT: AMD PAL Metadata: +# GNU-NEXT: Invalid AMD PAL Metadata + +--- !ELF +FileHeader: + Class: ELFCLASS64 + Data: ELFDATA2LSB + Type: ET_REL +Sections: + - Name: .note.nt_amd_hsa_code_object_version_0 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_CODE_OBJECT_VERSION + ## Desc size is too small (less than sizeof(CodeObjectVersion)). + Desc: '02000000' + - Name: .note.nt_amd_hsa_code_object_version_1 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_CODE_OBJECT_VERSION + ## Desc size is too large (more than sizeof(CodeObjectVersion)). + Desc: '020000000200000002000000' + - Name: .note.nt_amd_hsa_hsail_0 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_HSAIL + ## Desc size is too small (less than sizeof(HSAILProperties)). + Desc: '02000000010000000102' + - Name: .note.nt_amd_hsa_hsail_1 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_HSAIL + ## Desc size is too large (more than sizeof(HSAILProperties)). + Desc: '0200000001000000010202000000010200000001' + - Name: .note.nt_amd_hsa_isa_version_0 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_ISA_VERSION + ## Desc size is too small (less than sizeof(IsaVersion)). + Desc: '0400070008000000' + - Name: .note.nt_amd_hsa_isa_version_1 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_ISA_VERSION + ## Desc size is too small (less than sizeof(IsaVersion) + + ## VendorNameSize + ArchitectureNameSize). + Desc: '04000700080000000000000002000000414d4400414d44' + - Name: .note.nt_amd_hsa_isa_version_2 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_ISA_VERSION + ## VendorNameSize is 0. + Desc: '00000700080000000000000002000000414d4400414d44' + - Name: .note.nt_amd_hsa_isa_version_3 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_ISA_VERSION + ## ArchitectureNameSize is 0. + Desc: '04000000080000000000000002000000414d4400414d44' + - Name: .note.nt_amd_pal_metadata + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_PAL_METADATA + ## Desc size is not a multiple of sizeof(PALMetadata). + Desc: '02000000' diff --git a/llvm/test/tools/llvm-readobj/ELF/note-amd-invalid-v3.test b/llvm/test/tools/llvm-readobj/ELF/note-amd-invalid-v3.test new file mode 100644 index 0000000000000..b6713377d7ef4 --- /dev/null +++ b/llvm/test/tools/llvm-readobj/ELF/note-amd-invalid-v3.test @@ -0,0 +1,51 @@ +## This test is checking the handling of invalid/malformed note entries for +## AMDGPU code object v3. + +# RUN: yaml2obj %s -o %t.o +# RUN: llvm-readobj --notes %t.o | FileCheck %s --match-full-lines --check-prefix=LLVM +# RUN: llvm-readelf --notes %t.o | FileCheck %s --match-full-lines --check-prefix=GNU + +# LLVM: Notes [ +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amdgpu_metadata +# LLVM-NEXT: Offset: 0x40 +# LLVM-NEXT: Size: 0x28 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMDGPU +# LLVM-NEXT: Data size: 0x11 +# LLVM-NEXT: Type: NT_AMDGPU_METADATA (AMDGPU Metadata) +# LLVM-NEXT: AMDGPU Metadata: Invalid AMDGPU Metadata +# LLVM-NEXT: --- +# LLVM-NEXT: 0: 0 +# LLVM-NEXT: amdhsa.kernels: +# LLVM-NEXT: - 0 +# LLVM-NEXT: ... +# LLVM-EMPTY: +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: ] + +# GNU: Displaying notes found in: .note.nt_amdgpu_metadata +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMDGPU 0x00000011 NT_AMDGPU_METADATA (AMDGPU Metadata) +# GNU-NEXT: AMDGPU Metadata: +# GNU-NEXT: Invalid AMDGPU Metadata +# GNU-NEXT: --- +# GNU-NEXT: 0: 0 +# GNU-NEXT: amdhsa.kernels: +# GNU-NEXT: - 0 +# GNU-NEXT: ... + +--- !ELF +FileHeader: + Class: ELFCLASS64 + Data: ELFDATA2LSB + Type: ET_REL +Sections: + - Name: .note.nt_amdgpu_metadata + Type: SHT_NOTE + Notes: + - Name: AMDGPU + Type: NT_AMDGPU_METADATA + ## Desc contains 'amdhsa.kernels' without valid entries. + Desc: '82ae616d646873612e6b65726e656c7391' diff --git a/llvm/test/tools/llvm-readobj/ELF/note-amd-valid-v2.test b/llvm/test/tools/llvm-readobj/ELF/note-amd-valid-v2.test new file mode 100644 index 0000000000000..688bbd50994c3 --- /dev/null +++ b/llvm/test/tools/llvm-readobj/ELF/note-amd-valid-v2.test @@ -0,0 +1,202 @@ +## This test is checking the handling of valid note entries for AMDGPU code +## object v2. + +# RUN: yaml2obj %s -o %t.o +# RUN: llvm-readobj --notes %t.o | FileCheck %s --match-full-lines --check-prefix=LLVM +# RUN: llvm-readelf --notes %t.o | FileCheck %s --match-full-lines --check-prefix=GNU + +# LLVM: Notes [ +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_code_object_version +# LLVM-NEXT: Offset: 0x40 +# LLVM-NEXT: Size: 0x18 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x8 +# LLVM-NEXT: Type: NT_AMD_HSA_CODE_OBJECT_VERSION (AMD HSA Code Object Version) +# LLVM-NEXT: AMD HSA Code Object Version: [Major: 2, Minor: 1] +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_hsail +# LLVM-NEXT: Offset: 0x58 +# LLVM-NEXT: Size: 0x1C +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0xC +# LLVM-NEXT: Type: NT_AMD_HSA_HSAIL (AMD HSA HSAIL Properties) +# LLVM-NEXT: AMD HSA HSAIL Properties: [HSAIL Major: 2, HSAIL Minor: 1, Profile: 1, Machine Model: 2, Default Float Round: 3] +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_isa_version +# LLVM-NEXT: Offset: 0x74 +# LLVM-NEXT: Size: 0x2C +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x1B +# LLVM-NEXT: Type: NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# LLVM-NEXT: AMD HSA ISA Version: [Vendor: AMD, Architecture: AMDGPU, Major: 8, Minor: 0, Stepping: 2] +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_metadata_0 +# LLVM-NEXT: Offset: 0xA0 +# LLVM-NEXT: Size: 0x10 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x0 +# LLVM-NEXT: Type: NT_AMD_HSA_METADATA (AMD HSA Metadata) +# LLVM-NEXT: AMD HSA Metadata: +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_metadata_1 +# LLVM-NEXT: Offset: 0xB0 +# LLVM-NEXT: Size: 0x18 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x6 +# LLVM-NEXT: Type: NT_AMD_HSA_METADATA (AMD HSA Metadata) +# LLVM-NEXT: AMD HSA Metadata: abcde +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_isa_name_0 +# LLVM-NEXT: Offset: 0xC8 +# LLVM-NEXT: Size: 0x10 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x0 +# LLVM-NEXT: Type: NT_AMD_HSA_ISA_NAME (AMD HSA ISA Name) +# LLVM-NEXT: AMD HSA ISA Name: +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_hsa_isa_name_1 +# LLVM-NEXT: Offset: 0xD8 +# LLVM-NEXT: Size: 0x18 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x6 +# LLVM-NEXT: Type: NT_AMD_HSA_ISA_NAME (AMD HSA ISA Name) +# LLVM-NEXT: AMD HSA ISA Name: abcdef +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: NoteSection { +# LLVM-NEXT: Name: .note.nt_amd_pal_metadata +# LLVM-NEXT: Offset: 0xF0 +# LLVM-NEXT: Size: 0x28 +# LLVM-NEXT: Note { +# LLVM-NEXT: Owner: AMD +# LLVM-NEXT: Data size: 0x18 +# LLVM-NEXT: Type: NT_AMD_PAL_METADATA (AMD PAL Metadata) +# LLVM-NEXT: AMD PAL Metadata: [2: 1][4: 2][8: 4] +# LLVM-NEXT: } +# LLVM-NEXT: } +# LLVM-NEXT: ] + +# GNU: Displaying notes found in: .note.nt_amd_hsa_code_object_version +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000008 NT_AMD_HSA_CODE_OBJECT_VERSION (AMD HSA Code Object Version) +# GNU-NEXT: AMD HSA Code Object Version: +# GNU-NEXT: [Major: 2, Minor: 1] +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_hsail +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x0000000c NT_AMD_HSA_HSAIL (AMD HSA HSAIL Properties) +# GNU-NEXT: AMD HSA HSAIL Properties: +# GNU-NEXT: [HSAIL Major: 2, HSAIL Minor: 1, Profile: 1, Machine Model: 2, Default Float Round: 3] +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_isa_version +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x0000001b NT_AMD_HSA_ISA_VERSION (AMD HSA ISA Version) +# GNU-NEXT: AMD HSA ISA Version: +# GNU-NEXT: [Vendor: AMD, Architecture: AMDGPU, Major: 8, Minor: 0, Stepping: 2] +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_metadata_0 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000000 NT_AMD_HSA_METADATA (AMD HSA Metadata) +# GNU-NEXT: AMD HSA Metadata: +# GNU-NEXT: {{^ $}} +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_metadata_1 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000006 NT_AMD_HSA_METADATA (AMD HSA Metadata) +# GNU-NEXT: AMD HSA Metadata: +# GNU-NEXT: abcde +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_isa_name_0 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000000 NT_AMD_HSA_ISA_NAME (AMD HSA ISA Name) +# GNU-NEXT: AMD HSA ISA Name: +# GNU-NEXT: {{^ $}} +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_hsa_isa_name_1 +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000006 NT_AMD_HSA_ISA_NAME (AMD HSA ISA Name) +# GNU-NEXT: AMD HSA ISA Name: +# GNU-NEXT: abcdef +# GNU-EMPTY: +# GNU-NEXT: Displaying notes found in: .note.nt_amd_pal_metadata +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMD 0x00000018 NT_AMD_PAL_METADATA (AMD PAL Metadata) +# GNU-NEXT: AMD PAL Metadata: +# GNU-NEXT: [2: 1][4: 2][8: 4] + +--- !ELF +FileHeader: + Class: ELFCLASS64 + Data: ELFDATA2LSB + Type: ET_REL +Sections: + - Name: .note.nt_amd_hsa_code_object_version + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_CODE_OBJECT_VERSION + Desc: '0200000001000000' + - Name: .note.nt_amd_hsa_hsail + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_HSAIL + Desc: '020000000100000001020300' + - Name: .note.nt_amd_hsa_isa_version + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_ISA_VERSION + Desc: '04000700080000000000000002000000414d4400414d4447505500' + - Name: .note.nt_amd_hsa_metadata_0 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_METADATA + ## Check empty desc. + Desc: '' + - Name: .note.nt_amd_hsa_metadata_1 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_METADATA + Desc: '616263646566' + - Name: .note.nt_amd_hsa_isa_name_0 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_ISA_NAME + ## Check empty desc. + Desc: '' + - Name: .note.nt_amd_hsa_isa_name_1 + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_HSA_ISA_NAME + Desc: '616263646566' + - Name: .note.nt_amd_pal_metadata + Type: SHT_NOTE + Notes: + - Name: AMD + Type: NT_AMD_PAL_METADATA + Desc: '020000000100000004000000020000000800000004000000' diff --git a/llvm/test/tools/llvm-readobj/ELF/note-amd-valid-v3.s b/llvm/test/tools/llvm-readobj/ELF/note-amd-valid-v3.s new file mode 100644 index 0000000000000..3cf51d7214a8f --- /dev/null +++ b/llvm/test/tools/llvm-readobj/ELF/note-amd-valid-v3.s @@ -0,0 +1,73 @@ +## This test is checking the handling of valid note entries for AMDGPU code +## object v3. + +# RUN: llvm-mc -triple=amdgcn-amd-amdhsa -mcpu=gfx900 -filetype=obj < %s | llvm-readobj --notes - | FileCheck %s --match-full-lines --check-prefix=LLVM +# RUN: llvm-mc -triple=amdgcn-amd-amdhsa -mcpu=gfx900 -filetype=obj < %s | llvm-readelf --notes - | FileCheck %s --match-full-lines --check-prefix=GNU + +#LLVM: Notes [ +#LLVM-NEXT: NoteSection { +#LLVM-NEXT: Name: .note +#LLVM-NEXT: Offset: 0x40 +#LLVM-NEXT: Size: 0x110 +#LLVM-NEXT: Note { +#LLVM-NEXT: Owner: AMDGPU +#LLVM-NEXT: Data size: 0xFC +#LLVM-NEXT: Type: NT_AMDGPU_METADATA (AMDGPU Metadata) +#LLVM-NEXT: AMDGPU Metadata: --- +#LLVM-NEXT: amdhsa.kernels: +#LLVM-NEXT: - .group_segment_fixed_size: 16 +#LLVM-NEXT: .kernarg_segment_align: 64 +#LLVM-NEXT: .kernarg_segment_size: 8 +#LLVM-NEXT: .max_flat_workgroup_size: 256 +#LLVM-NEXT: .name: test_kernel +#LLVM-NEXT: .private_segment_fixed_size: 32 +#LLVM-NEXT: .sgpr_count: 14 +#LLVM-NEXT: .symbol: 'test_kernel@kd' +#LLVM-NEXT: .vgpr_count: 40 +#LLVM-NEXT: .wavefront_size: 128 +#LLVM-NEXT: amdhsa.version: +#LLVM-NEXT: - 1 +#LLVM-NEXT: - 0 +#LLVM-NEXT: ... +#LLVM-EMPTY: +#LLVM-NEXT: } +#LLVM-NEXT: } +#LLVM-NEXT: ] + +# GNU: Displaying notes found in: .note +# GNU-NEXT: Owner Data size Description +# GNU-NEXT: AMDGPU 0x000000fc NT_AMDGPU_METADATA (AMDGPU Metadata) +# GNU-NEXT: AMDGPU Metadata: +# GNU-NEXT: --- +# GNU-NEXT: amdhsa.kernels: +# GNU-NEXT: - .group_segment_fixed_size: 16 +# GNU-NEXT: .kernarg_segment_align: 64 +# GNU-NEXT: .kernarg_segment_size: 8 +# GNU-NEXT: .max_flat_workgroup_size: 256 +# GNU-NEXT: .name: test_kernel +# GNU-NEXT: .private_segment_fixed_size: 32 +# GNU-NEXT: .sgpr_count: 14 +# GNU-NEXT: .symbol: 'test_kernel@kd' +# GNU-NEXT: .vgpr_count: 40 +# GNU-NEXT: .wavefront_size: 128 +# GNU-NEXT: amdhsa.version: +# GNU-NEXT: - 1 +# GNU-NEXT: - 0 +# GNU-NEXT: ... + +.amdgpu_metadata + amdhsa.version: + - 1 + - 0 + amdhsa.kernels: + - .name: test_kernel + .symbol: test_kernel@kd + .group_segment_fixed_size: 16 + .kernarg_segment_align: 64 + .kernarg_segment_size: 8 + .max_flat_workgroup_size: 256 + .private_segment_fixed_size: 32 + .sgpr_count: 14 + .vgpr_count: 40 + .wavefront_size: 128 +.end_amdgpu_metadata diff --git a/llvm/tools/llvm-readobj/ELFDumper.cpp b/llvm/tools/llvm-readobj/ELFDumper.cpp index 45c61a61b0284..264641d865674 100644 --- a/llvm/tools/llvm-readobj/ELFDumper.cpp +++ b/llvm/tools/llvm-readobj/ELFDumper.cpp @@ -5055,9 +5055,10 @@ static AMDNote getAMDNote(uint32_t NoteType, ArrayRef Desc) { raw_string_ostream StrOS(HSAILPropetiesString); StrOS << "[HSAIL Major: " << Properties->HSAILMajorVersion << ", HSAIL Minor: " << Properties->HSAILMinorVersion - << ", Profile: " << Properties->Profile - << ", Machine Model: " << Properties->MachineModel - << ", Default Float Round: " << Properties->DefaultFloatRound << "]"; + << ", Profile: " << uint32_t(Properties->Profile) + << ", Machine Model: " << uint32_t(Properties->MachineModel) + << ", Default Float Round: " + << uint32_t(Properties->DefaultFloatRound) << "]"; return {"AMD HSA HSAIL Properties", HSAILPropetiesString}; } case ELF::NT_AMD_HSA_ISA_VERSION: { @@ -5105,10 +5106,12 @@ static AMDNote getAMDNote(uint32_t NoteType, ArrayRef Desc) { uint32_t Key; uint32_t Value; }; + if (Desc.size() % sizeof(PALMetadata) != 0) + return {"AMD PAL Metadata", "Invalid AMD PAL Metadata"}; auto Isa = reinterpret_cast(Desc.data()); std::string MetadataString; raw_string_ostream StrOS(MetadataString); - for (size_t I = 0, E = Desc.size() / sizeof(PALMetadata); I < E; ++E) { + for (size_t I = 0, E = Desc.size() / sizeof(PALMetadata); I < E; ++I) { StrOS << "[" << Isa[I].Key << ": " << Isa[I].Value << "]"; } return {"AMD PAL Metadata", MetadataString};