1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
| # RUN: yaml2obj %s > %t.o
# RUN: llvm-readobj --notes %t.o | FileCheck %s --check-prefix=LLVM
# RUN: llvm-readelf --notes %t.o | FileCheck %s --check-prefix=GNU
# GNU: Displaying notes found
# GNU-NEXT: Owner Data size Description
# GNU-NEXT: AMDGPU 0x000000e6 NT_AMDGPU_METADATA (AMDGPU Metadata)
# GNU-NEXT: AMDGPU Metadata:
# GNU-NEXT: ---
# GNU-NEXT: amdhsa.kernels:
# GNU-NEXT: - .group_segment_fixed_size: 2
# GNU-NEXT: .kernarg_segment_align: 4
# GNU-NEXT: .kernarg_segment_size: 1
# GNU-NEXT: .max_flat_workgroup_size: 8
# GNU-NEXT: .name: foo
# GNU-NEXT: .private_segment_fixed_size: 3
# GNU-NEXT: .sgpr_count: 6
# GNU-NEXT: .symbol: foo
# GNU-NEXT: .vgpr_count: 7
# GNU-NEXT: .wavefront_size: 5
# GNU-NEXT: amdhsa.version:
# GNU-NEXT: - 1
# GNU-NEXT: - 0
# GNU-NEXT: ...
# LLVM: Notes [
# LLVM-NEXT: NoteSection {
# LLVM-NEXT: Offset:
# LLVM-NEXT: Size:
# LLVM-NEXT: Note {
# LLVM-NEXT: Owner: AMDGPU
# LLVM-NEXT: Data size: 0xE6
# LLVM-NEXT: Type: NT_AMDGPU_METADATA (AMDGPU Metadata)
# LLVM-NEXT: AMDGPU Metadata: ---
# LLVM-NEXT: amdhsa.kernels:
# LLVM-NEXT: - .group_segment_fixed_size: 2
# LLVM-NEXT: .kernarg_segment_align: 4
# LLVM-NEXT: .kernarg_segment_size: 1
# LLVM-NEXT: .max_flat_workgroup_size: 8
# LLVM-NEXT: .name: foo
# LLVM-NEXT: .private_segment_fixed_size: 3
# LLVM-NEXT: .sgpr_count: 6
# LLVM-NEXT: .symbol: foo
# LLVM-NEXT: .vgpr_count: 7
# LLVM-NEXT: .wavefront_size: 5
# LLVM-NEXT: amdhsa.version:
# LLVM-NEXT: - 1
# LLVM-NEXT: - 0
# LLVM-NEXT: ...
# LLVM-EMPTY:
# LLVM-NEXT: }
# LLVM-NEXT: }
# LLVM-NEXT: ]
## Use yaml2obj instead of llvm-mc for more test portability. This was
## generated by grabbing section data from an object built via:
# $ llvm-mc -filetype=obj -triple amdgcn-amd-amdhsa -mattr=+code-object-v3 %s -o %t.o
## On input:
# .amdgpu_metadata
# amdhsa.version:
# - 1
# - 0
# amdhsa.kernels:
# - .name: foo
# .symbol: foo
# .kernarg_segment_size: 1
# .group_segment_fixed_size: 2
# .private_segment_fixed_size: 3
# .kernarg_segment_align: 4
# .wavefront_size: 5
# .sgpr_count: 6
# .vgpr_count: 7
# .max_flat_workgroup_size: 8
# .end_amdgpu_metadata
--- !ELF
FileHeader:
Class: ELFCLASS64
Data: ELFDATA2LSB
Type: ET_REL
Machine: EM_X86_64
Sections:
- Name: .note.foo
Type: SHT_NOTE
Content: 07000000E600000020000000414D44475055000082AE616D646873612E6B65726E656C73918AB92E67726F75705F7365676D656E745F66697865645F73697A6502B62E6B65726E6172675F7365676D656E745F616C69676E04B52E6B65726E6172675F7365676D656E745F73697A6501B82E6D61785F666C61745F776F726B67726F75705F73697A6508A52E6E616D65A3666F6FBB2E707269766174655F7365676D656E745F66697865645F73697A6503AB2E736770725F636F756E7406A72E73796D626F6CA3666F6FAB2E766770725F636F756E7407AF2E7761766566726F6E745F73697A6505AE616D646873612E76657273696F6E9201000000
|