* [PATCH 0/7] Add AMDGCN support to readelf
@ 2022-03-15 19:42 Simon Marchi
2022-03-15 19:42 ` [PATCH 1/7] bfd: add AMDGCN architecture Simon Marchi
` (7 more replies)
0 siblings, 8 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:42 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
This goal of this series is to add to readelf the capability to inspect
ELF files of the AMDGCN architecture. The AMDGCN architecture is an AMD
GPU architecture that can be used to execute compute kernelsi. ELF
files produced for that architecture (using the HIP compiler [3])
typically don't exist by themselves, but are bundled inside the ELF file
of the host program that will submit the kernels for execution on the
GPU. But other than that, they are regular ELF files.
The documentation for the ELF file format for the architecture can be
found here [1]. I refer to it a few times throughout the patches.
Our end goal is to upstream the GDB port for AMDGCN. But the readelf
bits are relatively simple, independent of GDB and useful on their own,
which is why I submit them first.
I uploaded this series to a users branch for convenience [4].
Since it may not be trivial for those trying for the first time, here
are instructions on how to install the ROCm stack and generate a binary
that can be used to test what's included in this series. Is it based on
the instructions found here [2]. I suggest using Ubuntu 20.04, since
that's what I tested with. This can all be done in a Docker container
to avoid polluting the system (it ends up installed a lot of packages).
- Install the amdgpu-install tool
$ wget https://repo.radeon.com/amdgpu-install/21.50/ubuntu/focal/amdgpu-install_21.50.50000-1_all.deb
$ sudo apt-get install ./amdgpu-install_21.50.50000-1_all.deb
- Install the ROCm packages
$ sudo amdgpu-install --usecase=rocm
- Checkout the users/simark/amdgpu-readelf binutils-gdb branch
$ git checkout users/simark/amdgpu-readelf
- Configure and build
$ ./configure --target=amdgcn-amd-amdhsa --disable-gas
$ make all-binutils
- Write a test source file with
#include "hip/hip_runtime.h"
__global__ void the_kernel() { __builtin_amdgcn_s_sleep(1); }
int main() { hipLaunchKernelGGL(the_kernel, dim3(1), dim3(1), 0, 0); }
- Compile the test source file
$ hipcc test.cpp
- Extract the bundled AMDGCN ELF file from the host ELF file (adjust
the URI accordingly, based on the output of roc-obj-ls)
$ /opt/rocm/bin/roc-obj-ls a.out
1 host-x86_64-unknown-linux file://a.out#offset=8192&size=0
1 hipv4-amdgcn-amd-amdhsa--gfx803 file://a.out#offset=8192&size=9536
$ /opt/rocm/bin/roc-obj-extract "file://a.out#offset=8192&size=9536"
- Run readelf on the extracted binary
$ ./binutils/readelf --header a.out-offset8192-size9536.co
[1] https://llvm.org/docs/AMDGPUUsage.html#elf-code-object
[2] https://github.com/RadeonOpenCompute/ROCm/blob/fac29ca466c7362a659ca4cd53fc2abc798a3c90/ROCm_Installation_Guide%20v5.0.pdf
[3] https://github.com/ROCm-Developer-Tools/HIP
[4] https://sourceware.org/git/?p=binutils-gdb.git;a=shortlog;h=refs/heads/users/simark/amdgpu-readelf
Simon Marchi (7):
bfd: add AMDGCN architecture
opcodes: handle bfd_amdgcn_arch in configure script
binutils/readelf: handle AMDGPU OS ABIs
binutils/readelf: decode AMDGPU-specific e_flags
binutils/readelf: handle NT_AMDGPU_METADATA note name
binutils/readelf: build against msgpack, dump NT_AMDGPU_METADATA note
contents
binutils/readelf: handle AMDGPU relocation types
bfd/Makefile.am | 4 +
bfd/Makefile.in | 6 +
bfd/archures.c | 15 ++
bfd/bfd-in2.h | 13 ++
bfd/config.bfd | 5 +
bfd/configure | 1 +
bfd/configure.ac | 1 +
bfd/cpu-amdgcn.c | 59 +++++++
bfd/elf-bfd.h | 1 +
bfd/elf64-amdgcn.c | 80 +++++++++
bfd/targets.c | 5 +
binutils/Makefile.am | 14 +-
binutils/Makefile.in | 11 +-
binutils/config.in | 3 +
binutils/configure | 137 ++++++++++++++-
binutils/configure.ac | 17 ++
binutils/readelf.c | 379 ++++++++++++++++++++++++++++++++++++++++--
include/elf/amdgpu.h | 115 +++++++++++++
include/elf/common.h | 3 +
opcodes/configure | 1 +
opcodes/configure.ac | 1 +
21 files changed, 852 insertions(+), 19 deletions(-)
create mode 100644 bfd/cpu-amdgcn.c
create mode 100644 bfd/elf64-amdgcn.c
create mode 100644 include/elf/amdgpu.h
base-commit: af481f01297a61af30f3f3dbaa11e6636c587dcb
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [PATCH 1/7] bfd: add AMDGCN architecture
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
@ 2022-03-15 19:42 ` Simon Marchi
2022-03-15 19:42 ` [PATCH 2/7] opcodes: handle bfd_amdgcn_arch in configure script Simon Marchi
` (6 subsequent siblings)
7 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:42 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
From: Simon Marchi <simon.marchi@efficios.com>
Add support for the AMDGCN architecture to BFD.
This is the bare minimum to get
$ ./configure --target=amdgcn-hsa-amdhsa --disable-gas
$ make all-binutils
working later in this series.
The specific AMDGCN models added here are a bit arbitrary, based on
what we intend to initially support in GDB. This list will need to be
updated in the future anyway. The complete up-to-date list of existing
AMDGPU models can be found here:
https://llvm.org/docs/AMDGPUUsage.html#processors
The ELF format for this architecture is documented here:
https://llvm.org/docs/AMDGPUUsage.html#elf-code-object
The flags for the "HSA" OS ABI are properly versioned and documented on
that page. But the NONE, PAL and MESA3D OS ABIs are not well documented
nor versioned. Taking a peek at the LLVM source code, we see that they
encode their flags the same way as HSA v3. For example, for PAL:
https://github.com/llvm/llvm-project/blob/c8b614cd74a92d85936aed5ac7c642af75ffdc29/llvm/lib/Target/AMDGPU/MCTargetDesc/AMDGPUTargetStreamer.cpp#L601
So at least, we know that all AMDGPU objects (of which AMDGCN objects
are a subset of) at the time of writing encode the specific GPU model in
the EF_AMDGPU_MACH field of e_flags.
bfd/ChangeLog:
* Makefile.am (ALL_MACHINES, ALL_MACHINES_CFILES):
Add cpu-amdgcn.c.
(BFD64_BACKENDS): Add elf64-amdgcn.lo.
(BFD64_BACKENDS_CFILES): Add elf64-amdgcn.c.
* Makefile.in: Re-generate.
* cpu-amdgcn.c: New.
* elf64-amdgcn.c: New.
* archures.c (bfd_architecture): Add bfd_arch_amdgcn and related
mach defines.
(bfd_amdgcn_arch): New.
(bfd_archures_list): Add bfd_amdgcn_arch.
* bfd-in2.h: Re-generate.
* config.bfd: Handle amdgcn* target.
* configure.ac: Handle amdgcn_elf64_le_vec.
* configure: Re-generate.
* elf-bfd.h (elf_target_id): Add AMDGCN_ELF_DATA.
* targets.c (amdgcn_elf64_le_vec): New.
(_bfd_target_vector): Add amdgcn_elf64_le_vec.
include/ChangeLog:
* elf/amdgpu.h: New.
* elf/common.h (ELFOSABI_AMDGPU_HSA): Add.
Change-Id: I969f7b14960797e88891c308749a6e341eece5b2
---
bfd/Makefile.am | 4 +++
bfd/Makefile.in | 6 ++++
bfd/archures.c | 15 +++++++++
bfd/bfd-in2.h | 13 +++++++
bfd/config.bfd | 5 +++
bfd/configure | 1 +
bfd/configure.ac | 1 +
bfd/cpu-amdgcn.c | 59 ++++++++++++++++++++++++++++++++
bfd/elf-bfd.h | 1 +
bfd/elf64-amdgcn.c | 80 ++++++++++++++++++++++++++++++++++++++++++++
bfd/targets.c | 5 +++
include/elf/amdgpu.h | 36 ++++++++++++++++++++
include/elf/common.h | 1 +
13 files changed, 227 insertions(+)
create mode 100644 bfd/cpu-amdgcn.c
create mode 100644 bfd/elf64-amdgcn.c
create mode 100644 include/elf/amdgpu.h
diff --git a/bfd/Makefile.am b/bfd/Makefile.am
index 0f0138408ce7..b9a3f8207aca 100644
--- a/bfd/Makefile.am
+++ b/bfd/Makefile.am
@@ -97,6 +97,7 @@ BFD64_LIBS_CFILES = archive64.c
ALL_MACHINES = \
cpu-aarch64.lo \
cpu-alpha.lo \
+ cpu-amdgcn.lo \
cpu-arc.lo \
cpu-arm.lo \
cpu-avr.lo \
@@ -182,6 +183,7 @@ ALL_MACHINES = \
ALL_MACHINES_CFILES = \
cpu-aarch64.c \
cpu-alpha.c \
+ cpu-amdgcn.c \
cpu-arc.c \
cpu-arm.c \
cpu-avr.c \
@@ -550,6 +552,7 @@ BFD64_BACKENDS = \
elf32-score.lo \
elf32-score7.lo \
elf64-alpha.lo \
+ elf64-amdgcn.lo \
elf64-gen.lo \
elf64-hppa.lo \
elf64-ia64.lo \
@@ -596,6 +599,7 @@ BFD64_BACKENDS_CFILES = \
elf32-score.c \
elf32-score7.c \
elf64-alpha.c \
+ elf64-amdgcn.c \
elf64-gen.c \
elf64-hppa.c \
elf64-ia64-vms.c \
diff --git a/bfd/Makefile.in b/bfd/Makefile.in
index b8e5ea0153fc..934dd4bc0664 100644
--- a/bfd/Makefile.in
+++ b/bfd/Makefile.in
@@ -564,6 +564,7 @@ BFD64_LIBS_CFILES = archive64.c
ALL_MACHINES = \
cpu-aarch64.lo \
cpu-alpha.lo \
+ cpu-amdgcn.lo \
cpu-arc.lo \
cpu-arm.lo \
cpu-avr.lo \
@@ -649,6 +650,7 @@ ALL_MACHINES = \
ALL_MACHINES_CFILES = \
cpu-aarch64.c \
cpu-alpha.c \
+ cpu-amdgcn.c \
cpu-arc.c \
cpu-arm.c \
cpu-avr.c \
@@ -1019,6 +1021,7 @@ BFD64_BACKENDS = \
elf32-score.lo \
elf32-score7.lo \
elf64-alpha.lo \
+ elf64-amdgcn.lo \
elf64-gen.lo \
elf64-hppa.lo \
elf64-ia64.lo \
@@ -1065,6 +1068,7 @@ BFD64_BACKENDS_CFILES = \
elf32-score.c \
elf32-score7.c \
elf64-alpha.c \
+ elf64-amdgcn.c \
elf64-gen.c \
elf64-hppa.c \
elf64-ia64-vms.c \
@@ -1480,6 +1484,7 @@ distclean-compile:
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/corefile.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/cpu-aarch64.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/cpu-alpha.Plo@am__quote@
+@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/cpu-amdgcn.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/cpu-arc.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/cpu-arm.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/cpu-avr.Plo@am__quote@
@@ -1646,6 +1651,7 @@ distclean-compile:
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/elf32.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/elf64-aarch64.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/elf64-alpha.Plo@am__quote@
+@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/elf64-amdgcn.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/elf64-bpf.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/elf64-gen.Plo@am__quote@
@AMDEP_TRUE@@am__include@ @am__quote@./$(DEPDIR)/elf64-hppa.Plo@am__quote@
diff --git a/bfd/archures.c b/bfd/archures.c
index d19b5d7ee6ba..fac9fe82a086 100644
--- a/bfd/archures.c
+++ b/bfd/archures.c
@@ -559,6 +559,19 @@ DESCRIPTION
. bfd_arch_loongarch, {* LoongArch *}
.#define bfd_mach_loongarch32 1
.#define bfd_mach_loongarch64 2
+. bfd_arch_amdgcn, {* AMDGCN *}
+.#define bfd_mach_amdgcn_unknown 0x000
+.#define bfd_mach_amdgcn_gfx900 0x02c
+.#define bfd_mach_amdgcn_gfx904 0x02e
+.#define bfd_mach_amdgcn_gfx906 0x02f
+.#define bfd_mach_amdgcn_gfx908 0x030
+.#define bfd_mach_amdgcn_gfx90a 0x03f
+.#define bfd_mach_amdgcn_gfx1010 0x033
+.#define bfd_mach_amdgcn_gfx1011 0x034
+.#define bfd_mach_amdgcn_gfx1012 0x035
+.#define bfd_mach_amdgcn_gfx1030 0x036
+.#define bfd_mach_amdgcn_gfx1031 0x037
+.#define bfd_mach_amdgcn_gfx1032 0x038
. bfd_arch_last
. };
*/
@@ -614,6 +627,7 @@ DESCRIPTION
extern const bfd_arch_info_type bfd_aarch64_arch;
extern const bfd_arch_info_type bfd_alpha_arch;
+extern const bfd_arch_info_type bfd_amdgcn_arch;
extern const bfd_arch_info_type bfd_arc_arch;
extern const bfd_arch_info_type bfd_arm_arch;
extern const bfd_arch_info_type bfd_avr_arch;
@@ -704,6 +718,7 @@ static const bfd_arch_info_type * const bfd_archures_list[] =
#else
&bfd_aarch64_arch,
&bfd_alpha_arch,
+ &bfd_amdgcn_arch,
&bfd_arc_arch,
&bfd_arm_arch,
&bfd_avr_arch,
diff --git a/bfd/bfd-in2.h b/bfd/bfd-in2.h
index db41e7eb7fe3..c0b563aec02a 100644
--- a/bfd/bfd-in2.h
+++ b/bfd/bfd-in2.h
@@ -1938,6 +1938,19 @@ enum bfd_architecture
bfd_arch_loongarch, /* LoongArch */
#define bfd_mach_loongarch32 1
#define bfd_mach_loongarch64 2
+ bfd_arch_amdgcn, /* AMDGCN */
+#define bfd_mach_amdgcn_unknown 0x000
+#define bfd_mach_amdgcn_gfx900 0x02c
+#define bfd_mach_amdgcn_gfx904 0x02e
+#define bfd_mach_amdgcn_gfx906 0x02f
+#define bfd_mach_amdgcn_gfx908 0x030
+#define bfd_mach_amdgcn_gfx90a 0x03f
+#define bfd_mach_amdgcn_gfx1010 0x033
+#define bfd_mach_amdgcn_gfx1011 0x034
+#define bfd_mach_amdgcn_gfx1012 0x035
+#define bfd_mach_amdgcn_gfx1030 0x036
+#define bfd_mach_amdgcn_gfx1031 0x037
+#define bfd_mach_amdgcn_gfx1032 0x038
bfd_arch_last
};
diff --git a/bfd/config.bfd b/bfd/config.bfd
index 02f40d4fba22..872685cfb722 100644
--- a/bfd/config.bfd
+++ b/bfd/config.bfd
@@ -184,6 +184,7 @@ alpha*) targ_archs=bfd_alpha_arch ;;
am33_2.0*) targ_archs=bfd_mn10300_arch ;;
arc*) targ_archs=bfd_arc_arch ;;
arm*) targ_archs=bfd_arm_arch ;;
+amdgcn*) targ_archs=bfd_amdgcn_arch ;;
bfin*) targ_archs=bfd_bfin_arch ;;
c30*) targ_archs=bfd_tic30_arch ;;
c4x*) targ_archs=bfd_tic4x_arch ;;
@@ -321,6 +322,10 @@ case "${targ}" in
targ_defvec=alpha_ecoff_le_vec
want64=true
;;
+ amdgcn-*-*)
+ targ_defvec=amdgcn_elf64_le_vec
+ want64=true
+ ;;
ia64*-*-freebsd* | ia64*-*-netbsd* | ia64*-*-linux-* | ia64*-*-elf* | ia64*-*-kfreebsd*-gnu)
targ_defvec=ia64_elf64_le_vec
targ_selvecs="ia64_elf64_be_vec ia64_pei_vec"
diff --git a/bfd/configure b/bfd/configure
index 4502e52decb1..0ef4c206fb0a 100755
--- a/bfd/configure
+++ b/bfd/configure
@@ -13364,6 +13364,7 @@ do
alpha_vms_vec) tb="$tb vms-alpha.lo vms-misc.lo vms-lib.lo"; target_size=64 ;;
alpha_vms_lib_txt_vec) tb="$tb vms-lib.lo vms-misc.lo" ;;
am33_elf32_linux_vec) tb="$tb elf32-am33lin.lo elf32.lo $elf" ;;
+ amdgcn_elf64_le_vec) tb="$tb elf64-amdgcn.lo elf64.lo $elf"; target_size=64 ;;
aout0_be_vec) tb="$tb aout0.lo aout32.lo" ;;
aout64_vec) tb="$tb demo64.lo aout64.lo"; target_size=64 ;;
aout_vec) tb="$tb host-aout.lo aout32.lo" ;;
diff --git a/bfd/configure.ac b/bfd/configure.ac
index 07f2074770f8..9e873736792b 100644
--- a/bfd/configure.ac
+++ b/bfd/configure.ac
@@ -443,6 +443,7 @@ do
alpha_vms_vec) tb="$tb vms-alpha.lo vms-misc.lo vms-lib.lo"; target_size=64 ;;
alpha_vms_lib_txt_vec) tb="$tb vms-lib.lo vms-misc.lo" ;;
am33_elf32_linux_vec) tb="$tb elf32-am33lin.lo elf32.lo $elf" ;;
+ amdgcn_elf64_le_vec) tb="$tb elf64-amdgcn.lo elf64.lo $elf"; target_size=64 ;;
aout0_be_vec) tb="$tb aout0.lo aout32.lo" ;;
aout64_vec) tb="$tb demo64.lo aout64.lo"; target_size=64 ;;
aout_vec) tb="$tb host-aout.lo aout32.lo" ;;
diff --git a/bfd/cpu-amdgcn.c b/bfd/cpu-amdgcn.c
new file mode 100644
index 000000000000..ef51e660ed87
--- /dev/null
+++ b/bfd/cpu-amdgcn.c
@@ -0,0 +1,59 @@
+/* BFD support for the AMDGCN GPU architecture.
+
+ Copyright (C) 2019-2022 Free Software Foundation, Inc.
+
+ This file is part of BFD, the Binary File Descriptor library.
+
+ This program is free software; you can redistribute it and/or modify
+ it under the terms of the GNU General Public License as published by
+ the Free Software Foundation; either version 3 of the License, or
+ (at your option) any later version.
+
+ This program is distributed in the hope that it will be useful,
+ but WITHOUT ANY WARRANTY; without even the implied warranty of
+ MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
+ GNU General Public License for more details.
+
+ You should have received a copy of the GNU General Public License
+ along with this program. If not, see <http://www.gnu.org/licenses/>. */
+
+#include "sysdep.h"
+#include "bfd.h"
+#include "libbfd.h"
+
+#define N(MACHINE, PRINTABLE_NAME, DEFAULT, NEXT) \
+ { \
+ 32, /* 32 bits in a word */ \
+ 64, /* 64 bits in an address */ \
+ 8, /* 8 bits in a byte */ \
+ bfd_arch_amdgcn, \
+ MACHINE, \
+ "amdgcn", \
+ PRINTABLE_NAME, \
+ 3, /* section align power */ \
+ DEFAULT, \
+ bfd_default_compatible, \
+ bfd_default_scan, \
+ bfd_arch_default_fill, \
+ NEXT, \
+ 0 \
+ }
+
+#define NN(index) (&arch_info_struct[index])
+
+static const bfd_arch_info_type arch_info_struct[] =
+{
+ N (bfd_mach_amdgcn_gfx904, "amdgcn:gfx904", false, NN (1)),
+ N (bfd_mach_amdgcn_gfx906, "amdgcn:gfx906", false, NN (2)),
+ N (bfd_mach_amdgcn_gfx908, "amdgcn:gfx908", false, NN (3)),
+ N (bfd_mach_amdgcn_gfx90a, "amdgcn:gfx90a", false, NN (4)),
+ N (bfd_mach_amdgcn_gfx1010, "amdgcn:gfx1010", false, NN (5)),
+ N (bfd_mach_amdgcn_gfx1011, "amdgcn:gfx1011", false, NN (6)),
+ N (bfd_mach_amdgcn_gfx1012, "amdgcn:gfx1012", false, NN (7)),
+ N (bfd_mach_amdgcn_gfx1030, "amdgcn:gfx1030", false, NN (8)),
+ N (bfd_mach_amdgcn_gfx1031, "amdgcn:gfx1031", false, NN (9)),
+ N (bfd_mach_amdgcn_gfx1032, "amdgcn:gfx1032", false, NULL)
+};
+
+const bfd_arch_info_type bfd_amdgcn_arch =
+ N (bfd_mach_amdgcn_gfx900, "amdgcn:gfx900", true, NN (0));
diff --git a/bfd/elf-bfd.h b/bfd/elf-bfd.h
index 4904e1e0aa9f..5c3985f6e571 100644
--- a/bfd/elf-bfd.h
+++ b/bfd/elf-bfd.h
@@ -499,6 +499,7 @@ enum elf_target_id
{
AARCH64_ELF_DATA = 1,
ALPHA_ELF_DATA,
+ AMDGCN_ELF_DATA,
ARC_ELF_DATA,
ARM_ELF_DATA,
AVR_ELF_DATA,
diff --git a/bfd/elf64-amdgcn.c b/bfd/elf64-amdgcn.c
new file mode 100644
index 000000000000..691fa0a34c16
--- /dev/null
+++ b/bfd/elf64-amdgcn.c
@@ -0,0 +1,80 @@
+/* AMDGCN ELF support for BFD.
+
+ Copyright (C) 2019-2022 Free Software Foundation, Inc.
+
+ This file is part of BFD, the Binary File Descriptor library.
+
+ This program is free software; you can redistribute it and/or modify
+ it under the terms of the GNU General Public License as published by
+ the Free Software Foundation; either version 3 of the License, or
+ (at your option) any later version.
+
+ This program is distributed in the hope that it will be useful,
+ but WITHOUT ANY WARRANTY; without even the implied warranty of
+ MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
+ GNU General Public License for more details.
+
+ You should have received a copy of the GNU General Public License
+ along with this program. If not, see <http://www.gnu.org/licenses/>. */
+
+/* This file handles ELF files that are of the AMDGCN architecture. The
+ format is documented here:
+
+ https://llvm.org/docs/AMDGPUUsage.html#elf-code-object */
+
+#include "sysdep.h"
+#include "bfd.h"
+#include "libbfd.h"
+#include "elf-bfd.h"
+#include "elf/amdgpu.h"
+
+#include <string.h>
+
+static bool
+elf64_amdgcn_object_p (bfd *abfd)
+{
+ Elf_Internal_Ehdr *hdr = elf_elfheader (abfd);
+ unsigned int mach;
+ unsigned char osabi;
+ unsigned char osabi_version;
+
+ BFD_ASSERT (hdr->e_machine == EM_AMDGPU);
+
+ osabi = hdr->e_ident[EI_OSABI];
+ osabi_version = hdr->e_ident[EI_ABIVERSION];
+
+ /* Objects with OS ABI HSA version 2 encoded the GPU model differently (in a
+ note), but they are deprecated, so we don't need to support them. Reject
+ them specifically.
+
+ At the time of writing, all AMDGCN objects encode the specific GPU
+ model in the EF_AMDGPU_MACH field of e_flags. */
+ if (osabi == ELFOSABI_AMDGPU_HSA
+ && osabi_version < ELFABIVERSION_AMDGPU_HSA_V3)
+ return false;
+
+ mach = elf_elfheader (abfd)->e_flags & EF_AMDGPU_MACH;
+
+ /* Avoid matching non-AMDGCN AMDGPU objects (e.g. r600). */
+ if (mach < EF_AMDGPU_MACH_AMDGCN_MIN)
+ return false;
+
+ bfd_default_set_arch_mach (abfd, bfd_arch_amdgcn, mach);
+ return true;
+}
+
+
+#define TARGET_LITTLE_SYM amdgcn_elf64_le_vec
+#define TARGET_LITTLE_NAME "elf64-amdgcn"
+#define ELF_ARCH bfd_arch_amdgcn
+#define ELF_TARGET_ID AMDGCN_ELF_DATA
+#define ELF_MACHINE_CODE EM_AMDGPU
+#define ELF_MAXPAGESIZE 0x10000 /* 64KB */
+#define ELF_COMMONPAGESIZE 0x1000 /* 4KB */
+
+#define bfd_elf64_bfd_reloc_type_lookup bfd_default_reloc_type_lookup
+#define bfd_elf64_bfd_reloc_name_lookup _bfd_norelocs_bfd_reloc_name_lookup
+
+#define elf_backend_object_p elf64_amdgcn_object_p
+
+#include "elf64-target.h"
diff --git a/bfd/targets.c b/bfd/targets.c
index 18fec45f02a9..417743efc0ed 100644
--- a/bfd/targets.c
+++ b/bfd/targets.c
@@ -686,6 +686,7 @@ extern const bfd_target alpha_elf64_fbsd_vec;
extern const bfd_target alpha_vms_vec;
extern const bfd_target alpha_vms_lib_txt_vec;
extern const bfd_target am33_elf32_linux_vec;
+extern const bfd_target amdgcn_elf64_le_vec;
extern const bfd_target aout_vec;
extern const bfd_target arc_elf32_be_vec;
extern const bfd_target arc_elf32_le_vec;
@@ -997,6 +998,10 @@ static const bfd_target * const _bfd_target_vector[] =
&aarch64_pei_vec,
#endif
+#ifdef BFD64
+ &amdgcn_elf64_le_vec,
+#endif
+
#ifdef BFD64
&alpha_ecoff_le_vec,
&alpha_elf64_vec,
diff --git a/include/elf/amdgpu.h b/include/elf/amdgpu.h
new file mode 100644
index 000000000000..daa472e5b346
--- /dev/null
+++ b/include/elf/amdgpu.h
@@ -0,0 +1,36 @@
+/* AMDGPU ELF support for BFD.
+
+ Copyright (C) 2019-2021 Free Software Foundation, Inc.
+
+ This file is part of BFD, the Binary File Descriptor library.
+
+ This program is free software; you can redistribute it and/or modify
+ it under the terms of the GNU General Public License as published by
+ the Free Software Foundation; either version 3 of the License, or
+ (at your option) any later version.
+
+ This program is distributed in the hope that it will be useful,
+ but WITHOUT ANY WARRANTY; without even the implied warranty of
+ MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
+ GNU General Public License for more details.
+
+ You should have received a copy of the GNU General Public License
+ along with this program. If not, see <http://www.gnu.org/licenses/>. */
+
+#ifndef _ELF_AMDGPU_H
+#define _ELF_AMDGPU_H
+
+/* e_ident[EI_ABIVERSION] values, when e_ident[EI_OSABI] is
+ ELFOSABI_AMDGPU_HSA. */
+
+#define ELFABIVERSION_AMDGPU_HSA_V2 0
+#define ELFABIVERSION_AMDGPU_HSA_V3 1
+#define ELFABIVERSION_AMDGPU_HSA_V4 2
+#define ELFABIVERSION_AMDGPU_HSA_V5 3
+
+/* Processor selection mask for EF_AMDGPU_MACH_* values. */
+
+#define EF_AMDGPU_MACH 0x0ff
+#define EF_AMDGPU_MACH_AMDGCN_MIN 0x020
+
+#endif /* _ELF_AMDGPU_H */
diff --git a/include/elf/common.h b/include/elf/common.h
index def04c323d39..a1cace406a47 100644
--- a/include/elf/common.h
+++ b/include/elf/common.h
@@ -77,6 +77,7 @@
#define ELFOSABI_OPENVOS 18 /* Stratus Technologies OpenVOS */
#define ELFOSABI_C6000_ELFABI 64 /* Bare-metal TMS320C6000 */
+#define ELFOSABI_AMDGPU_HSA 64 /* AMD HSA Runtime */
#define ELFOSABI_C6000_LINUX 65 /* Linux TMS320C6000 */
#define ELFOSABI_ARM_FDPIC 65 /* ARM FDPIC */
#define ELFOSABI_ARM 97 /* ARM */
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [PATCH 2/7] opcodes: handle bfd_amdgcn_arch in configure script
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
2022-03-15 19:42 ` [PATCH 1/7] bfd: add AMDGCN architecture Simon Marchi
@ 2022-03-15 19:42 ` Simon Marchi
2022-03-15 19:42 ` [PATCH 3/7] binutils/readelf: handle AMDGPU OS ABIs Simon Marchi
` (5 subsequent siblings)
7 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:42 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
From: Simon Marchi <simon.marchi@efficios.com>
There isn't an actual opcodes implementation for the AMDGCN arch (yet),
this is just the bare minimum to get
$ ./configure --target=amdgcn-hsa-amdhsa --disable-gas
$ make all-binutils
working later in this series.
opcodes/ChangeLog:
* configure.ac: Handle bfd_amdgcn_arch.
* configure: Re-generate.
Change-Id: Ib7d7c5533a803ed8b2a293e9275f667ed781ce79
---
opcodes/configure | 1 +
opcodes/configure.ac | 1 +
2 files changed, 2 insertions(+)
diff --git a/opcodes/configure b/opcodes/configure
index 7036e708478e..c98c5bcd0d17 100755
--- a/opcodes/configure
+++ b/opcodes/configure
@@ -12491,6 +12491,7 @@ if test x${all_targets} = xfalse ; then
case "$arch" in
bfd_aarch64_arch) ta="$ta aarch64-asm.lo aarch64-dis.lo aarch64-opc.lo aarch64-asm-2.lo aarch64-dis-2.lo aarch64-opc-2.lo" ;;
bfd_alpha_arch) ta="$ta alpha-dis.lo alpha-opc.lo" ;;
+ bfd_amdgcn_arch) ;;
bfd_arc_arch) ta="$ta arc-dis.lo arc-opc.lo arc-ext.lo" ;;
bfd_arm_arch) ta="$ta arm-dis.lo" ;;
bfd_avr_arch) ta="$ta avr-dis.lo" ;;
diff --git a/opcodes/configure.ac b/opcodes/configure.ac
index e1dc8de27923..dc340c4d1106 100644
--- a/opcodes/configure.ac
+++ b/opcodes/configure.ac
@@ -263,6 +263,7 @@ if test x${all_targets} = xfalse ; then
case "$arch" in
bfd_aarch64_arch) ta="$ta aarch64-asm.lo aarch64-dis.lo aarch64-opc.lo aarch64-asm-2.lo aarch64-dis-2.lo aarch64-opc-2.lo" ;;
bfd_alpha_arch) ta="$ta alpha-dis.lo alpha-opc.lo" ;;
+ bfd_amdgcn_arch) ;;
bfd_arc_arch) ta="$ta arc-dis.lo arc-opc.lo arc-ext.lo" ;;
bfd_arm_arch) ta="$ta arm-dis.lo" ;;
bfd_avr_arch) ta="$ta avr-dis.lo" ;;
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [PATCH 3/7] binutils/readelf: handle AMDGPU OS ABIs
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
2022-03-15 19:42 ` [PATCH 1/7] bfd: add AMDGCN architecture Simon Marchi
2022-03-15 19:42 ` [PATCH 2/7] opcodes: handle bfd_amdgcn_arch in configure script Simon Marchi
@ 2022-03-15 19:42 ` Simon Marchi
2022-03-15 19:43 ` [PATCH 4/7] binutils/readelf: decode AMDGPU-specific e_flags Simon Marchi
` (4 subsequent siblings)
7 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:42 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
From: Simon Marchi <simon.marchi@efficios.com>
When the machine is EM_AMDGPU, handle the various OS ABIs described
here:
https://llvm.org/docs/AMDGPUUsage.html#header
For a binary with the HSA OS ABI, the change looks like:
- OS/ABI: <unknown: 40>
+ OS/ABI: AMD HSA
binutils/ChangeLog:
* readelf.c (get_osabi_name): Handle EM_AMDGPU OS ABIs.
include/ChangeLog:
* elf/common.h (ELFOSABI_AMDGPU_PAL, ELFOSABI_AMDGPU_MESA3D):
New.
Change-Id: I383590c390f7dc2fe0f902f50038735626d71863
---
binutils/readelf.c | 11 +++++++++++
include/elf/common.h | 2 ++
2 files changed, 13 insertions(+)
diff --git a/binutils/readelf.c b/binutils/readelf.c
index 8b46052c7ded..e8974aacec5e 100644
--- a/binutils/readelf.c
+++ b/binutils/readelf.c
@@ -4284,6 +4284,17 @@ get_osabi_name (Filedata * filedata, unsigned int osabi)
if (osabi >= 64)
switch (filedata->file_header.e_machine)
{
+ case EM_AMDGPU:
+ switch (osabi)
+ {
+ case ELFOSABI_AMDGPU_HSA: return "AMD HSA";
+ case ELFOSABI_AMDGPU_PAL: return "AMD PAL";
+ case ELFOSABI_AMDGPU_MESA3D: return "AMD Mesa3D";
+ default:
+ break;
+ }
+ break;
+
case EM_ARM:
switch (osabi)
{
diff --git a/include/elf/common.h b/include/elf/common.h
index a1cace406a47..70d63e3299c0 100644
--- a/include/elf/common.h
+++ b/include/elf/common.h
@@ -79,7 +79,9 @@
#define ELFOSABI_C6000_ELFABI 64 /* Bare-metal TMS320C6000 */
#define ELFOSABI_AMDGPU_HSA 64 /* AMD HSA Runtime */
#define ELFOSABI_C6000_LINUX 65 /* Linux TMS320C6000 */
+#define ELFOSABI_AMDGPU_PAL 65 /* AMD PAL Runtime */
#define ELFOSABI_ARM_FDPIC 65 /* ARM FDPIC */
+#define ELFOSABI_AMDGPU_MESA3D 66 /* AMD Mesa3D Runtime */
#define ELFOSABI_ARM 97 /* ARM */
#define ELFOSABI_STANDALONE 255 /* Standalone (embedded) application */
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [PATCH 4/7] binutils/readelf: decode AMDGPU-specific e_flags
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
` (2 preceding siblings ...)
2022-03-15 19:42 ` [PATCH 3/7] binutils/readelf: handle AMDGPU OS ABIs Simon Marchi
@ 2022-03-15 19:43 ` Simon Marchi
2022-03-15 19:43 ` [PATCH 5/7] binutils/readelf: handle NT_AMDGPU_METADATA note name Simon Marchi
` (3 subsequent siblings)
7 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:43 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
From: Simon Marchi <simon.marchi@efficios.com>
Decode and print the AMDGPU-specific fields of e_flags, as documented
here:
https://llvm.org/docs/AMDGPUUsage.html#header
That is:
- The specific GPU model
- Whether the xnack and sramecc features are enabled
The result looks like:
- Flags: 0x52f
+ Flags: 0x52f, gfx906, xnack any, sramecc any
The flags for the "HSA" OS ABI are properly versioned and documented on
that page. But the NONE, PAL and MESA3D OS ABIs are not well documented
nor versioned. Taking a peek at the LLVM source code, we see that they
encode their flags the same way as HSA v3. For example, for PAL:
https://github.com/llvm/llvm-project/blob/c8b614cd74a92d85936aed5ac7c642af75ffdc29/llvm/lib/Target/AMDGPU/MCTargetDesc/AMDGPUTargetStreamer.cpp#L601
So for those other OS ABIs, we read them the same as HSA v3.
binutils/ChangeLog:
* readelf.c: Include elf/amdgcn.h.
(decode_AMDGPU_machine_flags): New.
(get_machine_flags): Handle flags for EM_AMDGPU machine type.
include/ChangeLog:
* elf/amdgcn.h: Add EF_AMDGPU_MACH_AMDGCN_* and
EF_AMDGPU_FEATURE_* defines.
Change-Id: Ib5b94df7cae0719a22cf4e4fd0629330e9485c12
---
binutils/readelf.c | 152 +++++++++++++++++++++++++++++++++++++++++++
include/elf/amdgpu.h | 54 +++++++++++++++
2 files changed, 206 insertions(+)
diff --git a/binutils/readelf.c b/binutils/readelf.c
index e8974aacec5e..00b5e546c1e7 100644
--- a/binutils/readelf.c
+++ b/binutils/readelf.c
@@ -92,6 +92,7 @@
#include "elf/aarch64.h"
#include "elf/alpha.h"
+#include "elf/amdgpu.h"
#include "elf/arc.h"
#include "elf/arm.h"
#include "elf/avr.h"
@@ -3565,6 +3566,153 @@ decode_NDS32_machine_flags (unsigned e_flags, char buf[], size_t size)
r += snprintf (buf + r, size -r, ", L2C");
}
+static void
+decode_AMDGPU_machine_flags (Filedata *filedata, unsigned int e_flags,
+ char *buf)
+{
+ unsigned char *e_ident = filedata->file_header.e_ident;
+ unsigned char osabi = e_ident[EI_OSABI];
+ unsigned char abiversion = e_ident[EI_ABIVERSION];
+ unsigned int mach;
+
+ /* HSA OS ABI v2 used a different encoding, but we don't need to support it,
+ it has been deprecated for a while.
+
+ The PAL, MESA3D and NONE OS ABIs are not properly versioned, at the time
+ of writing, they use the same flags as HSA v3, so the code below uses that
+ assumption. */
+ if (osabi == ELFOSABI_AMDGPU_HSA && abiversion < ELFABIVERSION_AMDGPU_HSA_V3)
+ return;
+
+ mach = e_flags & EF_AMDGPU_MACH;
+ switch (mach)
+ {
+#define AMDGPU_CASE(code, string) \
+ case code: strcat (buf, ", " string); break;
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX600, "gfx600")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX601, "gfx601")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX700, "gfx700")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX701, "gfx701")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX702, "gfx702")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX703, "gfx703")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX704, "gfx704")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX801, "gfx801")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX802, "gfx802")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX803, "gfx803")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX810, "gfx810")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX900, "gfx900")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX902, "gfx902")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX904, "gfx904")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX906, "gfx906")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX908, "gfx908")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX909, "gfx909")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX90C, "gfx90c")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1010, "gfx1010")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1011, "gfx1011")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1012, "gfx1012")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1030, "gfx1030")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1031, "gfx1031")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1032, "gfx1032")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1033, "gfx1033")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX602, "gfx602")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX705, "gfx705")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX805, "gfx805")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1035, "gfx1035")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1034, "gfx1034")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX90A, "gfx90a")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX940, "gfx940")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1013, "gfx1013")
+ AMDGPU_CASE (EF_AMDGPU_MACH_AMDGCN_GFX1036, "gfx1036")
+ default:
+ sprintf (buf, _(", <unknown AMDGPU GPU type: %#x>"), mach);
+ break;
+#undef AMDGPU_CASE
+ }
+
+ buf += strlen (buf);
+ e_flags &= ~EF_AMDGPU_MACH;
+
+ if ((osabi == ELFOSABI_AMDGPU_HSA
+ && abiversion == ELFABIVERSION_AMDGPU_HSA_V3)
+ || osabi != ELFOSABI_AMDGPU_HSA)
+ {
+ /* For HSA v3 and other OS ABIs. */
+ if (e_flags & EF_AMDGPU_FEATURE_XNACK_V3)
+ {
+ strcat (buf, ", xnack on");
+ buf += strlen (buf);
+ e_flags &= ~EF_AMDGPU_FEATURE_XNACK_V3;
+ }
+
+ if (e_flags & EF_AMDGPU_FEATURE_SRAMECC_V3)
+ {
+ strcat (buf, ", sramecc on");
+ buf += strlen (buf);
+ e_flags &= ~EF_AMDGPU_FEATURE_SRAMECC_V3;
+ }
+ }
+ else
+ {
+ /* For HSA v4+. */
+ int xnack, sramecc;
+
+ xnack = e_flags & EF_AMDGPU_FEATURE_XNACK_V4;
+ switch (xnack)
+ {
+ case EF_AMDGPU_FEATURE_XNACK_UNSUPPORTED_V4:
+ break;
+
+ case EF_AMDGPU_FEATURE_XNACK_ANY_V4:
+ strcat (buf, ", xnack any");
+ break;
+
+ case EF_AMDGPU_FEATURE_XNACK_OFF_V4:
+ strcat (buf, ", xnack off");
+ break;
+
+ case EF_AMDGPU_FEATURE_XNACK_ON_V4:
+ strcat (buf, ", xnack on");
+ break;
+
+ default:
+ sprintf (buf, _(", <unknown xnack value: %#x>"), xnack);
+ break;
+ }
+
+ buf += strlen (buf);
+ e_flags &= ~EF_AMDGPU_FEATURE_XNACK_V4;
+
+ sramecc = e_flags & EF_AMDGPU_FEATURE_SRAMECC_V4;
+ switch (sramecc)
+ {
+ case EF_AMDGPU_FEATURE_SRAMECC_UNSUPPORTED_V4:
+ break;
+
+ case EF_AMDGPU_FEATURE_SRAMECC_ANY_V4:
+ strcat (buf, ", sramecc any");
+ break;
+
+ case EF_AMDGPU_FEATURE_SRAMECC_OFF_V4:
+ strcat (buf, ", sramecc off");
+ break;
+
+ case EF_AMDGPU_FEATURE_SRAMECC_ON_V4:
+ strcat (buf, ", sramecc on");
+ break;
+
+ default:
+ sprintf (buf, _(", <unknown sramecc value: %#x>"), sramecc);
+ break;
+ }
+
+ buf += strlen (buf);
+ e_flags &= ~EF_AMDGPU_FEATURE_SRAMECC_V4;
+ }
+
+ if (e_flags != 0)
+ sprintf (buf, _(", unknown flags bits: %#x"), e_flags);
+}
+
static char *
get_machine_flags (Filedata * filedata, unsigned e_flags, unsigned e_machine)
{
@@ -3717,6 +3865,10 @@ get_machine_flags (Filedata * filedata, unsigned e_flags, unsigned e_machine)
}
break;
+ case EM_AMDGPU:
+ decode_AMDGPU_machine_flags (filedata, e_flags, buf);
+ break;
+
case EM_CYGNUS_MEP:
switch (e_flags & EF_MEP_CPU_MASK)
{
diff --git a/include/elf/amdgpu.h b/include/elf/amdgpu.h
index daa472e5b346..005064fc264e 100644
--- a/include/elf/amdgpu.h
+++ b/include/elf/amdgpu.h
@@ -33,4 +33,58 @@
#define EF_AMDGPU_MACH 0x0ff
#define EF_AMDGPU_MACH_AMDGCN_MIN 0x020
+#define EF_AMDGPU_MACH_AMDGCN_GFX600 0x020
+#define EF_AMDGPU_MACH_AMDGCN_GFX601 0x021
+#define EF_AMDGPU_MACH_AMDGCN_GFX700 0x022
+#define EF_AMDGPU_MACH_AMDGCN_GFX701 0x023
+#define EF_AMDGPU_MACH_AMDGCN_GFX702 0x024
+#define EF_AMDGPU_MACH_AMDGCN_GFX703 0x025
+#define EF_AMDGPU_MACH_AMDGCN_GFX704 0x026
+#define EF_AMDGPU_MACH_AMDGCN_GFX801 0x028
+#define EF_AMDGPU_MACH_AMDGCN_GFX802 0x029
+#define EF_AMDGPU_MACH_AMDGCN_GFX803 0x02a
+#define EF_AMDGPU_MACH_AMDGCN_GFX810 0x02b
+#define EF_AMDGPU_MACH_AMDGCN_GFX900 0x02c
+#define EF_AMDGPU_MACH_AMDGCN_GFX902 0x02d
+#define EF_AMDGPU_MACH_AMDGCN_GFX904 0x02e
+#define EF_AMDGPU_MACH_AMDGCN_GFX906 0x02f
+#define EF_AMDGPU_MACH_AMDGCN_GFX908 0x030
+#define EF_AMDGPU_MACH_AMDGCN_GFX909 0x031
+#define EF_AMDGPU_MACH_AMDGCN_GFX90C 0x032
+#define EF_AMDGPU_MACH_AMDGCN_GFX1010 0x033
+#define EF_AMDGPU_MACH_AMDGCN_GFX1011 0x034
+#define EF_AMDGPU_MACH_AMDGCN_GFX1012 0x035
+#define EF_AMDGPU_MACH_AMDGCN_GFX1030 0x036
+#define EF_AMDGPU_MACH_AMDGCN_GFX1031 0x037
+#define EF_AMDGPU_MACH_AMDGCN_GFX1032 0x038
+#define EF_AMDGPU_MACH_AMDGCN_GFX1033 0x039
+#define EF_AMDGPU_MACH_AMDGCN_GFX602 0x03a
+#define EF_AMDGPU_MACH_AMDGCN_GFX705 0x03b
+#define EF_AMDGPU_MACH_AMDGCN_GFX805 0x03c
+#define EF_AMDGPU_MACH_AMDGCN_GFX1035 0x03d
+#define EF_AMDGPU_MACH_AMDGCN_GFX1034 0x03e
+#define EF_AMDGPU_MACH_AMDGCN_GFX90A 0x03f
+#define EF_AMDGPU_MACH_AMDGCN_GFX940 0x040
+#define EF_AMDGPU_MACH_AMDGCN_GFX1013 0x042
+#define EF_AMDGPU_MACH_AMDGCN_GFX1036 0x045
+
+/* Code object v3 machine flags. */
+
+#define EF_AMDGPU_FEATURE_XNACK_V3 0x100
+#define EF_AMDGPU_FEATURE_SRAMECC_V3 0x200
+
+/* Code object v4 (and later) machine flags. */
+
+#define EF_AMDGPU_FEATURE_XNACK_V4 0x300
+#define EF_AMDGPU_FEATURE_XNACK_UNSUPPORTED_V4 0x000
+#define EF_AMDGPU_FEATURE_XNACK_ANY_V4 0x100
+#define EF_AMDGPU_FEATURE_XNACK_OFF_V4 0x200
+#define EF_AMDGPU_FEATURE_XNACK_ON_V4 0x300
+
+#define EF_AMDGPU_FEATURE_SRAMECC_V4 0xc00
+#define EF_AMDGPU_FEATURE_SRAMECC_UNSUPPORTED_V4 0x000
+#define EF_AMDGPU_FEATURE_SRAMECC_ANY_V4 0x400
+#define EF_AMDGPU_FEATURE_SRAMECC_OFF_V4 0x800
+#define EF_AMDGPU_FEATURE_SRAMECC_ON_V4 0xc00
+
#endif /* _ELF_AMDGPU_H */
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [PATCH 5/7] binutils/readelf: handle NT_AMDGPU_METADATA note name
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
` (3 preceding siblings ...)
2022-03-15 19:43 ` [PATCH 4/7] binutils/readelf: decode AMDGPU-specific e_flags Simon Marchi
@ 2022-03-15 19:43 ` Simon Marchi
2022-03-15 19:43 ` [PATCH 6/7] binutils/readelf: build against msgpack, dump NT_AMDGPU_METADATA note contents Simon Marchi
` (2 subsequent siblings)
7 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:43 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
From: Simon Marchi <simon.marchi@efficios.com>
Handle the NT_AMDGPU_METADATA note, which is described here:
https://llvm.org/docs/AMDGPUUsage.html#code-object-v3-note-records
As of this patch, just print out the name, not the contents, which is in
the msgpack format.
binutils/ChangeLog:
* readelf.c (get_amdgpu_elf_note_type): New.
(process_note): Handle "AMDGPU" notes.
include/ChangeLog:
* elf/amdgcn.h (NT_AMDGPU_METADATA): New.
Change-Id: Id2dba2e2aeaa55ef7464fb35aee9c7d5f96ddb23
---
binutils/readelf.c | 20 ++++++++++++++++++++
include/elf/amdgpu.h | 4 ++++
2 files changed, 24 insertions(+)
diff --git a/binutils/readelf.c b/binutils/readelf.c
index 00b5e546c1e7..91515bdf0faa 100644
--- a/binutils/readelf.c
+++ b/binutils/readelf.c
@@ -19724,6 +19724,22 @@ decode_x86_compat_2_isa (unsigned int bitmask)
}
}
+static const char *
+get_amdgpu_elf_note_type (unsigned int e_type)
+{
+ switch (e_type)
+ {
+ case NT_AMDGPU_METADATA:
+ return _("NT_AMDGPU_METADATA (code object metadata)");
+ default:
+ {
+ static char buf[64];
+ snprintf (buf, sizeof (buf), _("Unknown note type: (0x%08x)"), e_type);
+ return buf;
+ }
+ }
+}
+
static void
decode_x86_isa (unsigned int bitmask)
{
@@ -21313,6 +21329,10 @@ process_note (Elf_Internal_Note * pnote,
/* GNU-specific object file notes. */
nt = get_gnu_elf_note_type (pnote->type);
+ else if (startswith (pnote->namedata, "AMDGPU"))
+ /* AMDGPU-specific object file notes. */
+ nt = get_amdgpu_elf_note_type (pnote->type);
+
else if (startswith (pnote->namedata, "FreeBSD"))
/* FreeBSD-specific core file notes. */
nt = get_freebsd_elfcore_note_type (filedata, pnote->type);
diff --git a/include/elf/amdgpu.h b/include/elf/amdgpu.h
index 005064fc264e..e3c90dc74e85 100644
--- a/include/elf/amdgpu.h
+++ b/include/elf/amdgpu.h
@@ -87,4 +87,8 @@
#define EF_AMDGPU_FEATURE_SRAMECC_OFF_V4 0x800
#define EF_AMDGPU_FEATURE_SRAMECC_ON_V4 0xc00
+/* Notes. */
+
+#define NT_AMDGPU_METADATA 32
+
#endif /* _ELF_AMDGPU_H */
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [PATCH 6/7] binutils/readelf: build against msgpack, dump NT_AMDGPU_METADATA note contents
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
` (4 preceding siblings ...)
2022-03-15 19:43 ` [PATCH 5/7] binutils/readelf: handle NT_AMDGPU_METADATA note name Simon Marchi
@ 2022-03-15 19:43 ` Simon Marchi
2022-03-15 19:43 ` [PATCH 7/7] binutils/readelf: handle AMDGPU relocation types Simon Marchi
2022-03-16 0:08 ` [PATCH 0/7] Add AMDGCN support to readelf Alan Modra
7 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:43 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
From: Simon Marchi <simon.marchi@efficios.com>
[This patch might be a bit more controversial, given its adds an
optional dependency on a third party library. If this is really a
problems, we can drop it from the series. The consequence being that
there will be no human readable output for the AMDGPU ELF notes.]
The AMDGPU HSA OS ABI (code object v3 and above) defines the
NT_AMDGPU_METADATA ELF note [1]. The content is a msgpack object
describing, among other things, the kernels present in the code object
and how to call them.
I think it would be useful for readelf to be able to display the content
of those notes. msgpack is a structured format, a bit like JSON, except
not text-based. It is therefore possible to dump the contents in
human-readable form without knowledge of the specific layout of the
note.
Add configury to binutils to optionally check for the msgpack C library
[2]. Add There is a new --with{,out}-msgpack configure flag, and the actual
library lookup is done using pkg-config.
If msgpack support is enabled, dumping a NT_AMDGPU_METADATA note looks
like:
$ readelf --notes amdgpu-code-object
Displaying notes found in: .note
Owner Data size Description
AMDGPU 0x0000040d NT_AMDGPU_METADATA (code object metadata)
{
"amdhsa.kernels": [
{
".args": [
{
".address_space": "global",
".name": "out.coerce",
".offset": 0,
".size": 8,
".value_kind": "global_buffer",
},
<snip>
If msgpack support is disabled, dump the contents as hex, as is done
with notes that are not handled in a special way. This allows one to
decode the contents manually (maybe using a command-line msgpack
decoder) if really needed.
[1] https://llvm.org/docs/AMDGPUUsage.html#code-object-metadata
[2] https://github.com/msgpack/msgpack-c/tree/c_master
binutils/ChangeLog:
* Makefile.am (readelf_CFLAGS): New.
(readelf_LDADD): Add MSGPACK_LIBS.
* Makefile.in: Re-generate.
* config.in: Re-generate.
* configure: Re-generate.
* configure.ac: Add --with-msgpack flag and check for msgpack
using pkg-config.
* readelf.c: Include msgpack.h if HAVE_MSGPACK.
(print_note_contents_hex): New.
(print_indents): New.
(dump_msgpack_obj): New.
(dump_msgpack): New.
(print_amdgpu_note): New.
(process_note): Handle NT_AMDGPU_METADATA note contents.
Use print_note_contents_hex.
Change-Id: Ia60a654e620bc32dfdb1bccd845594e2af328b84
---
binutils/Makefile.am | 14 ++-
binutils/Makefile.in | 11 ++-
binutils/config.in | 3 +
binutils/configure | 137 +++++++++++++++++++++++++++++-
binutils/configure.ac | 17 ++++
binutils/readelf.c | 193 +++++++++++++++++++++++++++++++++++++++---
6 files changed, 356 insertions(+), 19 deletions(-)
diff --git a/binutils/Makefile.am b/binutils/Makefile.am
index 5b13af29cc38..751fbacce127 100644
--- a/binutils/Makefile.am
+++ b/binutils/Makefile.am
@@ -256,7 +256,7 @@ objcopy_SOURCES = objcopy.c not-strip.c rename.c $(WRITE_DEBUG_SRCS) $(BULIBS)
strings_SOURCES = strings.c $(BULIBS)
readelf_SOURCES = readelf.c version.c unwind-ia64.c dwarf.c demanguse.c $(ELFLIBS)
-readelf_LDADD = $(LIBCTF_NOBFD) $(LIBINTL) $(LIBIBERTY) $(ZLIB) $(DEBUGINFOD_LIBS)
+readelf_LDADD = $(LIBCTF_NOBFD) $(LIBINTL) $(LIBIBERTY) $(ZLIB) $(DEBUGINFOD_LIBS) $(MSGPACK_LIBS)
elfedit_SOURCES = elfedit.c version.c $(ELFLIBS)
elfedit_LDADD = $(LIBINTL) $(LIBIBERTY)
@@ -381,6 +381,18 @@ endif
$(AM_V_CC)$(COMPILE) $(DEBUGINFOD_CFLAGS) -c -o $@ $(srcdir)/dwarf.c
endif
+readelf.@OBJEXT@: readelf.c
+if am__fastdepCC
+ $(AM_V_CC)$(COMPILE) -MT $@ -MD -MP -MF $(DEPDIR)/$*.Tpo $(MSGPACK_CFLAGS) -c -o $@ $(srcdir)/readelf.c
+ $(AM_V_at)mv -f $(DEPDIR)/$*.Tpo $(DEPDIR)/$*.Po
+else
+if AMDEP
+ source='readelf.c' object='$@' libtool=no @AMDEPBACKSLASH@
+ DEPDIR=$(DEPDIR) $(CCDEPMODE) $(depcomp) @AMDEPBACKSLASH@
+endif
+ $(AM_V_CC)$(COMPILE) $(MSGPACK_CFLAGS) -c -o $@ $(srcdir)/readelf.c
+endif
+
sysroff.@OBJEXT@: sysroff.c
if am__fastdepCC
$(AM_V_CC)$(COMPILE) -MT $@ -MD -MP -MF $(DEPDIR)/$*.Tpo -c -o $@ `test -f sysroff.c || echo $(srcdir)/`sysroff.c $(NO_WERROR)
diff --git a/binutils/Makefile.in b/binutils/Makefile.in
index 87f78628380b..78d32b350e38 100644
--- a/binutils/Makefile.in
+++ b/binutils/Makefile.in
@@ -537,6 +537,8 @@ MKDIR_P = @MKDIR_P@
MKINSTALLDIRS = @MKINSTALLDIRS@
MSGFMT = @MSGFMT@
MSGMERGE = @MSGMERGE@
+MSGPACK_CFLAGS = @MSGPACK_CFLAGS@
+MSGPACK_LIBS = @MSGPACK_LIBS@
NM = @NM@
NMEDIT = @NMEDIT@
NO_WERROR = @NO_WERROR@
@@ -785,7 +787,7 @@ size_SOURCES = size.c $(BULIBS)
objcopy_SOURCES = objcopy.c not-strip.c rename.c $(WRITE_DEBUG_SRCS) $(BULIBS)
strings_SOURCES = strings.c $(BULIBS)
readelf_SOURCES = readelf.c version.c unwind-ia64.c dwarf.c demanguse.c $(ELFLIBS)
-readelf_LDADD = $(LIBCTF_NOBFD) $(LIBINTL) $(LIBIBERTY) $(ZLIB) $(DEBUGINFOD_LIBS)
+readelf_LDADD = $(LIBCTF_NOBFD) $(LIBINTL) $(LIBIBERTY) $(ZLIB) $(DEBUGINFOD_LIBS) $(MSGPACK_LIBS)
elfedit_SOURCES = elfedit.c version.c $(ELFLIBS)
elfedit_LDADD = $(LIBINTL) $(LIBIBERTY)
strip_new_SOURCES = objcopy.c is-strip.c rename.c $(WRITE_DEBUG_SRCS) $(BULIBS)
@@ -1919,6 +1921,13 @@ dwarf.@OBJEXT@: dwarf.c
@AMDEP_TRUE@@am__fastdepCC_FALSE@ DEPDIR=$(DEPDIR) $(CCDEPMODE) $(depcomp) @AMDEPBACKSLASH@
@am__fastdepCC_FALSE@ $(AM_V_CC)$(COMPILE) $(DEBUGINFOD_CFLAGS) -c -o $@ $(srcdir)/dwarf.c
+readelf.@OBJEXT@: readelf.c
+@am__fastdepCC_TRUE@ $(AM_V_CC)$(COMPILE) -MT $@ -MD -MP -MF $(DEPDIR)/$*.Tpo $(MSGPACK_CFLAGS) -c -o $@ $(srcdir)/readelf.c
+@am__fastdepCC_TRUE@ $(AM_V_at)mv -f $(DEPDIR)/$*.Tpo $(DEPDIR)/$*.Po
+@AMDEP_TRUE@@am__fastdepCC_FALSE@ source='readelf.c' object='$@' libtool=no @AMDEPBACKSLASH@
+@AMDEP_TRUE@@am__fastdepCC_FALSE@ DEPDIR=$(DEPDIR) $(CCDEPMODE) $(depcomp) @AMDEPBACKSLASH@
+@am__fastdepCC_FALSE@ $(AM_V_CC)$(COMPILE) $(MSGPACK_CFLAGS) -c -o $@ $(srcdir)/readelf.c
+
sysroff.@OBJEXT@: sysroff.c
@am__fastdepCC_TRUE@ $(AM_V_CC)$(COMPILE) -MT $@ -MD -MP -MF $(DEPDIR)/$*.Tpo -c -o $@ `test -f sysroff.c || echo $(srcdir)/`sysroff.c $(NO_WERROR)
@am__fastdepCC_TRUE@ $(AM_V_at)mv -f $(DEPDIR)/$*.Tpo $(DEPDIR)/$*.Po
diff --git a/binutils/config.in b/binutils/config.in
index e6fa66fc41b6..81bd143140e0 100644
--- a/binutils/config.in
+++ b/binutils/config.in
@@ -104,6 +104,9 @@
/* Define to 1 if you have a working `mmap' system call. */
#undef HAVE_MMAP
+/* Define to 1 if msgpack is available. */
+#undef HAVE_MSGPACK
+
/* Define to 1 if you have the `sbrk' function. */
#undef HAVE_SBRK
diff --git a/binutils/configure b/binutils/configure
index 19d82badc638..f3ad831ad381 100755
--- a/binutils/configure
+++ b/binutils/configure
@@ -648,6 +648,8 @@ BUILD_DLLTOOL
BUILD_SRCONV
LTLIBICONV
LIBICONV
+MSGPACK_LIBS
+MSGPACK_CFLAGS
zlibinc
zlibdir
DEMANGLER_NAME
@@ -830,6 +832,7 @@ enable_build_warnings
enable_nls
enable_maintainer_mode
with_system_zlib
+with_msgpack
enable_rpath
with_libiconv_prefix
with_libiconv_type
@@ -849,7 +852,9 @@ PKG_CONFIG_LIBDIR
DEBUGINFOD_CFLAGS
DEBUGINFOD_LIBS
YACC
-YFLAGS'
+YFLAGS
+MSGPACK_CFLAGS
+MSGPACK_LIBS'
# Initialize some variables set by options.
@@ -1512,6 +1517,7 @@ Optional Packages:
--with-debuginfod Enable debuginfo lookups with debuginfod
(auto/yes/no)
--with-system-zlib use installed libz
+ --with-msgpack Enable msgpack support (auto/yes/no)
--with-gnu-ld assume the C compiler uses GNU ld default=no
--with-libiconv-prefix[=DIR] search for libiconv in DIR/include and DIR/lib
--without-libiconv-prefix don't search for libiconv in includedir and libdir
@@ -1541,6 +1547,10 @@ Some influential environment variables:
YFLAGS The list of arguments that will be passed by default to $YACC.
This script will default YFLAGS to the empty string to avoid a
default value of `-d' given by some make applications.
+ MSGPACK_CFLAGS
+ C compiler flags for MSGPACK, overriding pkg-config
+ MSGPACK_LIBS
+ linker flags for MSGPACK, overriding pkg-config
Use these variables to override the choices made by `configure' or to help
it to find libraries and programs with nonstandard names/locations.
@@ -10971,7 +10981,7 @@ else
lt_dlunknown=0; lt_dlno_uscore=1; lt_dlneed_uscore=2
lt_status=$lt_dlunknown
cat > conftest.$ac_ext <<_LT_EOF
-#line 10974 "configure"
+#line 10984 "configure"
#include "confdefs.h"
#if HAVE_DLFCN_H
@@ -11077,7 +11087,7 @@ else
lt_dlunknown=0; lt_dlno_uscore=1; lt_dlneed_uscore=2
lt_status=$lt_dlunknown
cat > conftest.$ac_ext <<_LT_EOF
-#line 11080 "configure"
+#line 11090 "configure"
#include "confdefs.h"
#if HAVE_DLFCN_H
@@ -13723,6 +13733,127 @@ $as_echo "#define USE_BINARY_FOPEN 1" >>confdefs.h
;;
esac
+# Support for the msgpack C library.
+
+# Check whether --with-msgpack was given.
+if test "${with_msgpack+set}" = set; then :
+ withval=$with_msgpack;
+else
+ with_msgpack=auto
+fi
+
+
+
+if test "$with_msgpack" != no; then
+
+pkg_failed=no
+{ $as_echo "$as_me:${as_lineno-$LINENO}: checking for msgpack" >&5
+$as_echo_n "checking for msgpack... " >&6; }
+
+if test -n "$MSGPACK_CFLAGS"; then
+ pkg_cv_MSGPACK_CFLAGS="$MSGPACK_CFLAGS"
+ elif test -n "$PKG_CONFIG"; then
+ if test -n "$PKG_CONFIG" && \
+ { { $as_echo "$as_me:${as_lineno-$LINENO}: \$PKG_CONFIG --exists --print-errors \"msgpack\""; } >&5
+ ($PKG_CONFIG --exists --print-errors "msgpack") 2>&5
+ ac_status=$?
+ $as_echo "$as_me:${as_lineno-$LINENO}: \$? = $ac_status" >&5
+ test $ac_status = 0; }; then
+ pkg_cv_MSGPACK_CFLAGS=`$PKG_CONFIG --cflags "msgpack" 2>/dev/null`
+ test "x$?" != "x0" && pkg_failed=yes
+else
+ pkg_failed=yes
+fi
+ else
+ pkg_failed=untried
+fi
+if test -n "$MSGPACK_LIBS"; then
+ pkg_cv_MSGPACK_LIBS="$MSGPACK_LIBS"
+ elif test -n "$PKG_CONFIG"; then
+ if test -n "$PKG_CONFIG" && \
+ { { $as_echo "$as_me:${as_lineno-$LINENO}: \$PKG_CONFIG --exists --print-errors \"msgpack\""; } >&5
+ ($PKG_CONFIG --exists --print-errors "msgpack") 2>&5
+ ac_status=$?
+ $as_echo "$as_me:${as_lineno-$LINENO}: \$? = $ac_status" >&5
+ test $ac_status = 0; }; then
+ pkg_cv_MSGPACK_LIBS=`$PKG_CONFIG --libs "msgpack" 2>/dev/null`
+ test "x$?" != "x0" && pkg_failed=yes
+else
+ pkg_failed=yes
+fi
+ else
+ pkg_failed=untried
+fi
+
+if test $pkg_failed = no; then
+ pkg_save_LDFLAGS="$LDFLAGS"
+ LDFLAGS="$LDFLAGS $pkg_cv_MSGPACK_LIBS"
+ cat confdefs.h - <<_ACEOF >conftest.$ac_ext
+/* end confdefs.h. */
+
+int
+main ()
+{
+
+ ;
+ return 0;
+}
+_ACEOF
+if ac_fn_c_try_link "$LINENO"; then :
+
+else
+ pkg_failed=yes
+fi
+rm -f core conftest.err conftest.$ac_objext \
+ conftest$ac_exeext conftest.$ac_ext
+ LDFLAGS=$pkg_save_LDFLAGS
+fi
+
+
+
+if test $pkg_failed = yes; then
+ { $as_echo "$as_me:${as_lineno-$LINENO}: result: no" >&5
+$as_echo "no" >&6; }
+
+if $PKG_CONFIG --atleast-pkgconfig-version 0.20; then
+ _pkg_short_errors_supported=yes
+else
+ _pkg_short_errors_supported=no
+fi
+ if test $_pkg_short_errors_supported = yes; then
+ MSGPACK_PKG_ERRORS=`$PKG_CONFIG --short-errors --print-errors --cflags --libs "msgpack" 2>&1`
+ else
+ MSGPACK_PKG_ERRORS=`$PKG_CONFIG --print-errors --cflags --libs "msgpack" 2>&1`
+ fi
+ # Put the nasty error message in config.log where it belongs
+ echo "$MSGPACK_PKG_ERRORS" >&5
+
+
+ if test "$with_msgpack" = yes; then
+ as_fn_error $? "--with-msgpack was given, but msgpack is missing or unusable." "$LINENO" 5
+ fi
+
+elif test $pkg_failed = untried; then
+ { $as_echo "$as_me:${as_lineno-$LINENO}: result: no" >&5
+$as_echo "no" >&6; }
+
+ if test "$with_msgpack" = yes; then
+ as_fn_error $? "--with-msgpack was given, but msgpack is missing or unusable." "$LINENO" 5
+ fi
+
+else
+ MSGPACK_CFLAGS=$pkg_cv_MSGPACK_CFLAGS
+ MSGPACK_LIBS=$pkg_cv_MSGPACK_LIBS
+ { $as_echo "$as_me:${as_lineno-$LINENO}: result: yes" >&5
+$as_echo "yes" >&6; }
+
+
+$as_echo "#define HAVE_MSGPACK 1" >>confdefs.h
+
+
+fi
+fi
+
# target-specific stuff:
# Canonicalize the secondary target names.
diff --git a/binutils/configure.ac b/binutils/configure.ac
index 7b70e7652435..e204a201a0a9 100644
--- a/binutils/configure.ac
+++ b/binutils/configure.ac
@@ -275,6 +275,23 @@ AM_ZLIB
BFD_BINARY_FOPEN
+# Support for the msgpack C library.
+AC_ARG_WITH([msgpack],
+ AC_HELP_STRING([--with-msgpack], [Enable msgpack support (auto/yes/no)]),
+ [],
+ [with_msgpack=auto])
+
+
+if test "$with_msgpack" != no; then
+ PKG_CHECK_MODULES(MSGPACK, msgpack, [
+ AC_DEFINE([HAVE_MSGPACK], [1], [Define to 1 if msgpack is available.])
+ ], [
+ if test "$with_msgpack" = yes; then
+ AC_MSG_ERROR([--with-msgpack was given, but msgpack is missing or unusable.])
+ fi
+ ])
+fi
+
# target-specific stuff:
# Canonicalize the secondary target names.
diff --git a/binutils/readelf.c b/binutils/readelf.c
index 91515bdf0faa..ff07112eb843 100644
--- a/binutils/readelf.c
+++ b/binutils/readelf.c
@@ -46,6 +46,10 @@
#include <zlib.h>
#include <wchar.h>
+#if defined HAVE_MSGPACK
+#include <msgpack.h>
+#endif
+
#if __GNUC__ >= 2
/* Define BFD64 here, even if our default architecture is 32 bit ELF
as this will allow us to read in and parse 64bit and 32bit ELF files.
@@ -21307,6 +21311,177 @@ print_gnu_build_attribute_name (Elf_Internal_Note * pnote)
return true;
}
+/* Print the contents of PNOTE as hex. */
+
+static void
+print_note_contents_hex (Elf_Internal_Note *pnote)
+{
+ if (pnote->descsz)
+ {
+ unsigned long i;
+
+ printf (_(" description data: "));
+ for (i = 0; i < pnote->descsz; i++)
+ printf ("%02x ", pnote->descdata[i] & 0xff);
+ if (!do_wide)
+ printf ("\n");
+ }
+
+ if (do_wide)
+ printf ("\n");
+}
+
+#if defined HAVE_MSGPACK
+
+static void
+print_indents (int n)
+{
+ printf (" ");
+
+ for (int i = 0; i < n; i++)
+ printf (" ");
+}
+
+/* Print OBJ in human-readable form. */
+
+static void
+dump_msgpack_obj (const msgpack_object *obj, int indent)
+{
+ switch (obj->type)
+ {
+ case MSGPACK_OBJECT_NIL:
+ printf ("(nil)");
+ break;
+
+ case MSGPACK_OBJECT_BOOLEAN:
+ printf ("%s", obj->via.boolean ? "true" : "false");
+ break;
+
+ case MSGPACK_OBJECT_POSITIVE_INTEGER:
+ printf ("%" PRIu64, obj->via.u64);
+ break;
+
+ case MSGPACK_OBJECT_NEGATIVE_INTEGER:
+ printf ("%" PRIi64, obj->via.i64);
+ break;
+
+ case MSGPACK_OBJECT_FLOAT32:
+ case MSGPACK_OBJECT_FLOAT64:
+ printf ("%f", obj->via.f64);
+ break;
+
+ case MSGPACK_OBJECT_STR:
+ printf ("\"%.*s\"", obj->via.str.size, obj->via.str.ptr);
+ break;
+
+ case MSGPACK_OBJECT_ARRAY:
+ {
+ const msgpack_object_array *array = &obj->via.array;
+
+ printf ("[\n");
+ ++indent;
+
+ for (uint32_t i = 0; i < array->size; ++i)
+ {
+ const msgpack_object *item = &array->ptr[i];
+
+ print_indents (indent);
+ dump_msgpack_obj (item, indent);
+ printf (",\n");
+ }
+
+ --indent;
+ print_indents (indent);
+ printf ("]");
+ break;
+ }
+ break;
+
+ case MSGPACK_OBJECT_MAP:
+ {
+ const msgpack_object_map *map = &obj->via.map;
+
+ printf ("{\n");
+ ++indent;
+
+ for (uint32_t i = 0; i < map->size; ++i)
+ {
+ const msgpack_object_kv *kv = &map->ptr[i];
+ const msgpack_object *key = &kv->key;
+ const msgpack_object *val = &kv->val;
+
+ print_indents (indent);
+ dump_msgpack_obj (key, indent);
+ printf (": ");
+ dump_msgpack_obj (val, indent);
+
+ printf (",\n");
+ }
+
+ --indent;
+ print_indents (indent);
+ printf ("}");
+
+ break;
+ }
+
+ case MSGPACK_OBJECT_BIN:
+ printf ("(bin)");
+ break;
+
+ case MSGPACK_OBJECT_EXT:
+ printf ("(ext)");
+ break;
+ }
+}
+
+static void
+dump_msgpack (const msgpack_unpacked *msg)
+{
+ print_indents (0);
+ dump_msgpack_obj (&msg->data, 0);
+ printf ("\n");
+}
+
+#endif /* defined HAVE_MSGPACK */
+
+static bool
+print_amdgpu_note (Elf_Internal_Note *pnote)
+{
+#if defined HAVE_MSGPACK
+ /* If msgpack is available, decode and dump the note's content. */
+ bool ret;
+ msgpack_unpacked msg;
+ msgpack_unpack_return msgpack_ret;
+
+ assert (pnote->type == NT_AMDGPU_METADATA);
+
+ msgpack_unpacked_init (&msg);
+ msgpack_ret = msgpack_unpack_next (&msg, pnote->descdata, pnote->descsz,
+ NULL);
+
+ switch (msgpack_ret)
+ {
+ case MSGPACK_UNPACK_SUCCESS:
+ dump_msgpack (&msg);
+ ret = true;
+ break;
+
+ default:
+ error (_("failed to unpack msgpack contents in NT_AMDGPU_METADATA note"));
+ ret = false;
+ break;
+ }
+
+ msgpack_unpacked_destroy (&msg);
+ return ret;
+#else
+ /* msgpack is not available, dump contents as hex. */
+ print_note_contents_hex (pnote);
+ return true;
+#endif
+}
+
/* Note that by the ELF standard, the name field is already null byte
terminated, and namesz includes the terminating null byte.
I.E. the value of namesz for the name "FSF" is 4.
@@ -21404,21 +21579,11 @@ process_note (Elf_Internal_Note * pnote,
&& (pnote->type == NT_GNU_BUILD_ATTRIBUTE_OPEN
|| pnote->type == NT_GNU_BUILD_ATTRIBUTE_FUNC))
return print_gnu_build_attribute_description (pnote, filedata);
+ else if (startswith (pnote->namedata, "AMDGPU")
+ && pnote->type == NT_AMDGPU_METADATA)
+ return print_amdgpu_note (pnote);
- if (pnote->descsz)
- {
- unsigned long i;
-
- printf (_(" description data: "));
- for (i = 0; i < pnote->descsz; i++)
- printf ("%02x ", pnote->descdata[i] & 0xff);
- if (!do_wide)
- printf ("\n");
- }
-
- if (do_wide)
- printf ("\n");
-
+ print_note_contents_hex (pnote);
return true;
}
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [PATCH 7/7] binutils/readelf: handle AMDGPU relocation types
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
` (5 preceding siblings ...)
2022-03-15 19:43 ` [PATCH 6/7] binutils/readelf: build against msgpack, dump NT_AMDGPU_METADATA note contents Simon Marchi
@ 2022-03-15 19:43 ` Simon Marchi
2022-03-16 0:08 ` [PATCH 0/7] Add AMDGCN support to readelf Alan Modra
7 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-15 19:43 UTC (permalink / raw)
To: binutils; +Cc: Simon Marchi
From: Simon Marchi <simon.marchi@efficios.com>
Make readelf recognize AMDGPU relocation types, as documented here:
https://llvm.org/docs/AMDGPUUsage.html#amdgpu-relocation-records
The user-visible change looks like:
-000000000004 000400000001 unrecognized: 1 0000000000000000 SCRATCH_RSRC_DWORD0
-00000000000c 000500000001 unrecognized: 1 0000000000000000 SCRATCH_RSRC_DWORD1
-000000000014 000600000007 unrecognized: 7 0000000000000000 global_var0
-00000000001c 000700000008 unrecognized: 8 0000000000000000 global_var1
-000000000024 000800000009 unrecognized: 9 0000000000000000 global_var2
-00000000002c 00090000000a unrecognized: a 0000000000000000 global_var3
-000000000034 000a0000000b unrecognized: b 0000000000000000 global_var4
+000000000004 000400000001 R_AMDGPU_ABS32_LO 0000000000000000 SCRATCH_RSRC_DWORD0
+00000000000c 000500000001 R_AMDGPU_ABS32_LO 0000000000000000 SCRATCH_RSRC_DWORD1
+000000000014 000600000007 R_AMDGPU_GOTPCREL 0000000000000000 global_var0
+00000000001c 000700000008 R_AMDGPU_GOTPCREL 0000000000000000 global_var1
+000000000024 000800000009 R_AMDGPU_GOTPCREL 0000000000000000 global_var2
+00000000002c 00090000000a R_AMDGPU_REL32_LO 0000000000000000 global_var3
+000000000034 000a0000000b R_AMDGPU_REL32_HI 0000000000000000 global_var4
binutils/ChangeLog:
* readelf.c (dump_relocations): Handle EM_AMDGPU.
include/ChangeLog:
* elf/amdgpu.h: Add relocation values.
Change-Id: I2ed4589f4cd37ea11ad2e0cb38d4b682271e1334
---
binutils/readelf.c | 3 +++
include/elf/amdgpu.h | 21 +++++++++++++++++++++
2 files changed, 24 insertions(+)
diff --git a/binutils/readelf.c b/binutils/readelf.c
index ff07112eb843..a2dbaa4722e3 100644
--- a/binutils/readelf.c
+++ b/binutils/readelf.c
@@ -1930,6 +1930,9 @@ dump_relocations (Filedata * filedata,
rtype = elf_loongarch_reloc_type (type);
break;
+ case EM_AMDGPU:
+ rtype = elf_amdgpu_reloc_type (type);
+ break;
}
if (rtype == NULL)
diff --git a/include/elf/amdgpu.h b/include/elf/amdgpu.h
index e3c90dc74e85..140e8996ad25 100644
--- a/include/elf/amdgpu.h
+++ b/include/elf/amdgpu.h
@@ -20,6 +20,8 @@
#ifndef _ELF_AMDGPU_H
#define _ELF_AMDGPU_H
+#include "elf/reloc-macros.h"
+
/* e_ident[EI_ABIVERSION] values, when e_ident[EI_OSABI] is
ELFOSABI_AMDGPU_HSA. */
@@ -91,4 +93,23 @@
#define NT_AMDGPU_METADATA 32
+/* Relocations. */
+
+START_RELOC_NUMBERS (elf_amdgpu_reloc_type)
+ RELOC_NUMBER (R_AMDGPU_NONE, 0)
+ RELOC_NUMBER (R_AMDGPU_ABS32_LO, 1)
+ RELOC_NUMBER (R_AMDGPU_ABS32_HI, 2)
+ RELOC_NUMBER (R_AMDGPU_ABS64, 3)
+ RELOC_NUMBER (R_AMDGPU_REL32, 4)
+ RELOC_NUMBER (R_AMDGPU_REL64, 5)
+ RELOC_NUMBER (R_AMDGPU_ABS32, 6)
+ RELOC_NUMBER (R_AMDGPU_GOTPCREL, 7)
+ RELOC_NUMBER (R_AMDGPU_GOTPCREL32_LO, 8)
+ RELOC_NUMBER (R_AMDGPU_GOTPCREL32_HI, 9)
+ RELOC_NUMBER (R_AMDGPU_REL32_LO, 10)
+ RELOC_NUMBER (R_AMDGPU_REL32_HI, 11)
+ RELOC_NUMBER (R_AMDGPU_RELATIVE64, 13)
+ RELOC_NUMBER (R_AMDGPU_REL16, 16)
+END_RELOC_NUMBERS (R_AMDGPU_max)
+
#endif /* _ELF_AMDGPU_H */
--
2.35.1
^ permalink raw reply [flat|nested] 10+ messages in thread
* Re: [PATCH 0/7] Add AMDGCN support to readelf
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
` (6 preceding siblings ...)
2022-03-15 19:43 ` [PATCH 7/7] binutils/readelf: handle AMDGPU relocation types Simon Marchi
@ 2022-03-16 0:08 ` Alan Modra
2022-03-16 13:02 ` Simon Marchi
7 siblings, 1 reply; 10+ messages in thread
From: Alan Modra @ 2022-03-16 0:08 UTC (permalink / raw)
To: Simon Marchi; +Cc: binutils
The series looks good to me.
--
Alan Modra
Australia Development Lab, IBM
^ permalink raw reply [flat|nested] 10+ messages in thread
* Re: [PATCH 0/7] Add AMDGCN support to readelf
2022-03-16 0:08 ` [PATCH 0/7] Add AMDGCN support to readelf Alan Modra
@ 2022-03-16 13:02 ` Simon Marchi
0 siblings, 0 replies; 10+ messages in thread
From: Simon Marchi @ 2022-03-16 13:02 UTC (permalink / raw)
To: Alan Modra; +Cc: binutils
On 2022-03-15 20:08, Alan Modra wrote:
> The series looks good to me.
>
Thanks, pushed.
Simon
^ permalink raw reply [flat|nested] 10+ messages in thread
end of thread, other threads:[~2022-03-16 13:02 UTC | newest]
Thread overview: 10+ messages (download: mbox.gz / follow: Atom feed)
-- links below jump to the message on this page --
2022-03-15 19:42 [PATCH 0/7] Add AMDGCN support to readelf Simon Marchi
2022-03-15 19:42 ` [PATCH 1/7] bfd: add AMDGCN architecture Simon Marchi
2022-03-15 19:42 ` [PATCH 2/7] opcodes: handle bfd_amdgcn_arch in configure script Simon Marchi
2022-03-15 19:42 ` [PATCH 3/7] binutils/readelf: handle AMDGPU OS ABIs Simon Marchi
2022-03-15 19:43 ` [PATCH 4/7] binutils/readelf: decode AMDGPU-specific e_flags Simon Marchi
2022-03-15 19:43 ` [PATCH 5/7] binutils/readelf: handle NT_AMDGPU_METADATA note name Simon Marchi
2022-03-15 19:43 ` [PATCH 6/7] binutils/readelf: build against msgpack, dump NT_AMDGPU_METADATA note contents Simon Marchi
2022-03-15 19:43 ` [PATCH 7/7] binutils/readelf: handle AMDGPU relocation types Simon Marchi
2022-03-16 0:08 ` [PATCH 0/7] Add AMDGCN support to readelf Alan Modra
2022-03-16 13:02 ` Simon Marchi
This is a public inbox, see mirroring instructions
for how to clone and mirror all data and code used for this inbox;
as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).