public inbox for gcc-bugs@sourceware.org
help / color / mirror / Atom feed
* [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
@ 2021-09-30 10:17 tschwinge at gcc dot gnu.org
2021-09-30 12:38 ` [Bug target/102544] " ams at gcc dot gnu.org
` (8 more replies)
0 siblings, 9 replies; 10+ messages in thread
From: tschwinge at gcc dot gnu.org @ 2021-09-30 10:17 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
Bug ID: 102544
Summary: GCN offloading not working for
'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
Product: gcc
Version: 11.0
Status: UNCONFIRMED
Keywords: openacc, openmp
Severity: normal
Priority: P3
Component: target
Assignee: unassigned at gcc dot gnu.org
Reporter: tschwinge at gcc dot gnu.org
CC: ams at gcc dot gnu.org, caplanr at predsci dot com,
jules at gcc dot gnu.org
Target Milestone: ---
Target: GCN
Ron Caplan (in CC) is always eager to try his computational codes with GCC (in
addition to other/proprietary/vendor-provided compilers), and I very much
appreciate him providing feedback of any kind, which I find is always
constructive.
He's recently reported that "using Ubuntu 20.04 with the latest rocm and a
Radeon VII", they got GCN offloading "to compile but when we run it we get:
'libgomp: target function wasn't mapped'".
He sent me 'rocm-smi' and 'rocminfo' output, and comparing his vs. that of our
amd_ryzen3 system, I find the following differences:
-/+ Agent [...]
*******
Name: gfx906
-/+ Uuid: [...]
- Marketing Name: Vega 20
+ Marketing Name: Vega 20 [Radeon VII]
Vendor Name: AMD
[...]
- Node: 1
+ Node: 2
Device Type: GPU
Cache Info:
L1: 16(0x10) KB
+ L2: 8192(0x2000) KB
Chip ID: 26287(0x66af)
Cacheline Size: 64(0x40)
-/+ Max Clock Freq. (MHz): [...]
- BDFID: 2816
+ BDFID: 33536
- Internal Node ID: 1
+ Internal Node ID: 2
Compute Unit: 60
SIMDs per CU: 4
Shader Engines: 4
@@ -128,7 +188,7 @@ Agent 2
Accessible by all: FALSE
ISA Info:
ISA 1
- Name: amdgcn-amd-amdhsa--gfx906:sramecc-:xnack-
+ Name: amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-
The very last item, I suppose, is the crucial one: 'sramecc-' vs. 'sramecc+' --
Andrew?
(a) How do we improve upon the run-time 'libgomp: target function wasn't
mapped' error reporting mode?
(b) What needs to be done to make this work?
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
@ 2021-09-30 12:38 ` ams at gcc dot gnu.org
2021-10-01 5:44 ` miko at predsci dot com
` (7 subsequent siblings)
8 siblings, 0 replies; 10+ messages in thread
From: ams at gcc dot gnu.org @ 2021-09-30 12:38 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #1 from Andrew Stubbs <ams at gcc dot gnu.org> ---
Please set "export GCN_DEBUG=1", try it again, and post the output.
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
2021-09-30 12:38 ` [Bug target/102544] " ams at gcc dot gnu.org
@ 2021-10-01 5:44 ` miko at predsci dot com
2021-10-01 10:54 ` ams at gcc dot gnu.org
` (6 subsequent siblings)
8 siblings, 0 replies; 10+ messages in thread
From: miko at predsci dot com @ 2021-10-01 5:44 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
miko at predsci dot com changed:
What |Removed |Added
----------------------------------------------------------------------------
CC| |miko at predsci dot com
--- Comment #2 from miko at predsci dot com ---
GCN debug: HSA run-time initialized for GCN
GCN debug: HSA_SYSTEM_INFO_ENDIANNESS: LITTLE
GCN debug: HSA_SYSTEM_INFO_EXTENSIONS: IMAGES
GCN debug: There are 1 GCN GPU devices.
GCN debug: HSA_AGENT_INFO_NAME: Intel(R) Xeon(R) CPU E5-2680 v3 @ 2.50GHz
GCN debug: HSA_AGENT_INFO_VENDOR_NAME: CPU
GCN debug: HSA_AGENT_INFO_MACHINE_MODEL: LARGE
GCN debug: HSA_AGENT_INFO_PROFILE: FULL
GCN debug: HSA_AGENT_INFO_DEVICE: CPU
GCN debug: HSA_AMD_AGENT_INFO_COMPUTE_UNIT_COUNT: 24
GCN debug: HSA_AGENT_INFO_WAVEFRONT_SIZE: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_SIZE: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_SIZE: 0
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_AGENT_INFO_NAME: Intel(R) Xeon(R) CPU E5-2680 v3 @ 2.50GHz
GCN debug: HSA_AGENT_INFO_VENDOR_NAME: CPU
GCN debug: HSA_AGENT_INFO_MACHINE_MODEL: LARGE
GCN debug: HSA_AGENT_INFO_PROFILE: FULL
GCN debug: HSA_AGENT_INFO_DEVICE: CPU
GCN debug: HSA_AMD_AGENT_INFO_COMPUTE_UNIT_COUNT: 24
GCN debug: HSA_AGENT_INFO_WAVEFRONT_SIZE: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_SIZE: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_SIZE: 0
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_AGENT_INFO_NAME: gfx906
GCN debug: HSA_AGENT_INFO_VENDOR_NAME: AMD
GCN debug: HSA_AGENT_INFO_MACHINE_MODEL: LARGE
GCN debug: HSA_AGENT_INFO_PROFILE: BASE
GCN debug: HSA_AGENT_INFO_DEVICE: GPU
GCN debug: HSA_AMD_AGENT_INFO_COMPUTE_UNIT_COUNT: 60
GCN debug: HSA_AGENT_INFO_WAVEFRONT_SIZE: 64
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_DIM: 67109888
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_SIZE: 1024
GCN debug: HSA_AGENT_INFO_GRID_MAX_DIM: 4294967295
GCN debug: HSA_AGENT_INFO_GRID_MAX_SIZE: 4294967295
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 17163091968
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 17163091968
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GROUP
GCN debug: HSA_REGION_INFO_SIZE: 65536
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 0
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 0
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
libgomp: target function wasn't mapped
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
2021-09-30 12:38 ` [Bug target/102544] " ams at gcc dot gnu.org
2021-10-01 5:44 ` miko at predsci dot com
@ 2021-10-01 10:54 ` ams at gcc dot gnu.org
2021-10-01 17:07 ` miko at predsci dot com
` (5 subsequent siblings)
8 siblings, 0 replies; 10+ messages in thread
From: ams at gcc dot gnu.org @ 2021-10-01 10:54 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #3 from Andrew Stubbs <ams at gcc dot gnu.org> ---
That output shows that we have the correct libgomp and rocm is installed and
working. Libgomp initialized the GCN plugin, but did not attempt to initialize
the device (the next message in the output should have been "Selected kernel
arguments memory region", or at least a GCN error message).
Instead we have a target-independent libgomp error. Presumably the kernel
metadata is malformed, somehow?
I think we need a testcase to debug this further, preferably reduced to be as
simple as possible.
Perhaps it would be a good idea to start with a minimal toy example and see if
that works on the device.
#include <openacc.h>
#include <stdio.h>
int main ()
{
int v = 1;
#pragma acc parallel copy(v)
{
if (acc_on_device(acc_device_host))
v = -1; // error
else {
v = 2; // success
}
}
printf ("v is %d\n", v);
return v;
}
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
` (2 preceding siblings ...)
2021-10-01 10:54 ` ams at gcc dot gnu.org
@ 2021-10-01 17:07 ` miko at predsci dot com
2021-10-01 17:21 ` ams at gcc dot gnu.org
` (4 subsequent siblings)
8 siblings, 0 replies; 10+ messages in thread
From: miko at predsci dot com @ 2021-10-01 17:07 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #4 from miko at predsci dot com ---
The result is:
v is -1
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
` (3 preceding siblings ...)
2021-10-01 17:07 ` miko at predsci dot com
@ 2021-10-01 17:21 ` ams at gcc dot gnu.org
2021-10-01 18:44 ` miko at predsci dot com
` (3 subsequent siblings)
8 siblings, 0 replies; 10+ messages in thread
From: ams at gcc dot gnu.org @ 2021-10-01 17:21 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #5 from Andrew Stubbs <ams at gcc dot gnu.org> ---
Sorry, I should have said to compile with -fopenacc.
If you did do that, please post the GCN_DEBUG output.
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
` (4 preceding siblings ...)
2021-10-01 17:21 ` ams at gcc dot gnu.org
@ 2021-10-01 18:44 ` miko at predsci dot com
2021-10-01 18:48 ` caplanr at predsci dot com
` (2 subsequent siblings)
8 siblings, 0 replies; 10+ messages in thread
From: miko at predsci dot com @ 2021-10-01 18:44 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #6 from miko at predsci dot com ---
Q: /home/mikostul/test_gfortran $ ./Small_test_F
GCN debug: HSA run-time initialized for GCN
GCN debug: HSA_SYSTEM_INFO_ENDIANNESS: LITTLE
GCN debug: HSA_SYSTEM_INFO_EXTENSIONS: IMAGES
GCN debug: There are 1 GCN GPU devices.
GCN debug: HSA_AGENT_INFO_NAME: Intel(R) Xeon(R) CPU E5-2680 v3 @ 2.50GHz
GCN debug: HSA_AGENT_INFO_VENDOR_NAME: CPU
GCN debug: HSA_AGENT_INFO_MACHINE_MODEL: LARGE
GCN debug: HSA_AGENT_INFO_PROFILE: FULL
GCN debug: HSA_AGENT_INFO_DEVICE: CPU
GCN debug: HSA_AMD_AGENT_INFO_COMPUTE_UNIT_COUNT: 24
GCN debug: HSA_AGENT_INFO_WAVEFRONT_SIZE: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_SIZE: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_SIZE: 0
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_AGENT_INFO_NAME: Intel(R) Xeon(R) CPU E5-2680 v3 @ 2.50GHz
GCN debug: HSA_AGENT_INFO_VENDOR_NAME: CPU
GCN debug: HSA_AGENT_INFO_MACHINE_MODEL: LARGE
GCN debug: HSA_AGENT_INFO_PROFILE: FULL
GCN debug: HSA_AGENT_INFO_DEVICE: CPU
GCN debug: HSA_AMD_AGENT_INFO_COMPUTE_UNIT_COUNT: 24
GCN debug: HSA_AGENT_INFO_WAVEFRONT_SIZE: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_SIZE: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_DIM: 0
GCN debug: HSA_AGENT_INFO_GRID_MAX_SIZE: 0
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_AGENT_INFO_NAME: gfx906
GCN debug: HSA_AGENT_INFO_VENDOR_NAME: AMD
GCN debug: HSA_AGENT_INFO_MACHINE_MODEL: LARGE
GCN debug: HSA_AGENT_INFO_PROFILE: BASE
GCN debug: HSA_AGENT_INFO_DEVICE: GPU
GCN debug: HSA_AMD_AGENT_INFO_COMPUTE_UNIT_COUNT: 60
GCN debug: HSA_AGENT_INFO_WAVEFRONT_SIZE: 64
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_DIM: 67109888
GCN debug: HSA_AGENT_INFO_WORKGROUP_MAX_SIZE: 1024
GCN debug: HSA_AGENT_INFO_GRID_MAX_DIM: 4294967295
GCN debug: HSA_AGENT_INFO_GRID_MAX_SIZE: 4294967295
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 17163091968
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 17163091968
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GROUP
GCN debug: HSA_REGION_INFO_SIZE: 65536
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 0
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 0
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: KERNARG
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: FINE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33595305984
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
GCN debug: HSA_REGION_INFO_SEGMENT: GLOBAL
GCN debug: HSA_REGION_INFO_GLOBAL_FLAGS: COARSE_GRAINED
GCN debug: HSA_REGION_INFO_SIZE: 33816289280
GCN debug: HSA_REGION_INFO_ALLOC_MAX_SIZE: 134823190528
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALLOWED: 1
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_GRANULE: 4096
GCN debug: HSA_REGION_INFO_RUNTIME_ALLOC_ALIGNMENT: 4096
libgomp: target function wasn't mapped
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
` (5 preceding siblings ...)
2021-10-01 18:44 ` miko at predsci dot com
@ 2021-10-01 18:48 ` caplanr at predsci dot com
2021-10-04 12:07 ` ams at gcc dot gnu.org
2021-10-04 16:52 ` miko at predsci dot com
8 siblings, 0 replies; 10+ messages in thread
From: caplanr at predsci dot com @ 2021-10-01 18:48 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #7 from Ron <caplanr at predsci dot com> ---
That last case was using a fortran version of the test code as follows:
program SMALL_TEST_F
use openacc
implicit none
integer :: v
v = 0
!$acc parallel copy(v)
if (acc_on_device(ACC_DEVICE_HOST)) then
v = -1
else
v = 2
end if
!$acc end parallel
print *,'v is',v
end program
compiled with gfortran 10 with the flags:
-g -fopenacc -foffload=amdgcn-amdhsa="-march=gfx906" -foffload=-lgfortran
-foffload=-lm
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
` (6 preceding siblings ...)
2021-10-01 18:48 ` caplanr at predsci dot com
@ 2021-10-04 12:07 ` ams at gcc dot gnu.org
2021-10-04 16:52 ` miko at predsci dot com
8 siblings, 0 replies; 10+ messages in thread
From: ams at gcc dot gnu.org @ 2021-10-04 12:07 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #8 from Andrew Stubbs <ams at gcc dot gnu.org> ---
Did you get the C version to return anything other than "-1"? (The expected
result is "2".)
I'm still trying to determine if the device is compatible, but the mapping
problem looks like a different issue.
Your code works fine on my device using a somewhat more recent GCC build. (I
can't install that exact toolchain right now.)
^ permalink raw reply [flat|nested] 10+ messages in thread
* [Bug target/102544] GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-'
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
` (7 preceding siblings ...)
2021-10-04 12:07 ` ams at gcc dot gnu.org
@ 2021-10-04 16:52 ` miko at predsci dot com
8 siblings, 0 replies; 10+ messages in thread
From: miko at predsci dot com @ 2021-10-04 16:52 UTC (permalink / raw)
To: gcc-bugs
https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102544
--- Comment #9 from miko at predsci dot com ---
Both the C and Fortran versions returned "-1".
^ permalink raw reply [flat|nested] 10+ messages in thread
end of thread, other threads:[~2021-10-04 16:52 UTC | newest]
Thread overview: 10+ messages (download: mbox.gz / follow: Atom feed)
-- links below jump to the message on this page --
2021-09-30 10:17 [Bug target/102544] New: GCN offloading not working for 'amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-' tschwinge at gcc dot gnu.org
2021-09-30 12:38 ` [Bug target/102544] " ams at gcc dot gnu.org
2021-10-01 5:44 ` miko at predsci dot com
2021-10-01 10:54 ` ams at gcc dot gnu.org
2021-10-01 17:07 ` miko at predsci dot com
2021-10-01 17:21 ` ams at gcc dot gnu.org
2021-10-01 18:44 ` miko at predsci dot com
2021-10-01 18:48 ` caplanr at predsci dot com
2021-10-04 12:07 ` ams at gcc dot gnu.org
2021-10-04 16:52 ` miko at predsci dot com
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).