From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: from esa1.mentor.iphmx.com (esa1.mentor.iphmx.com [68.232.129.153]) by sourceware.org (Postfix) with ESMTPS id 261853857C7C for ; Thu, 13 Jan 2022 12:22:28 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.1 sourceware.org 261853857C7C Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=mentor.com IronPort-SDR: zQiK9XhXYruz6LjUFsOvjawED7sTgkKI9G/3lVbMRJiAgAQbjfzpK9MYIpXAbGBGq2TuQfiOkK 9AwNkfqJPo4H7CSV2VUn6ivx5fqzuzmCJ4/y+sBaDGlwzZQxJelUJ+Gl5tOSCRREGcLaHBEFra sOc8MwYTyL4uKaZTYVnHjC8yAddZnHvhFkOQJNo0KOdW2ngTZZ0/J3icnNR5e5zmFdqF1c0x8L 9RTCxXDHQatyKguXlVzj+YlMVltkwNFqDEs/7i3EIbX7MF475RwkDyb74ih1B60wub6Mdssb4U QE53oPqCiThahl0+qpEjgQL2 X-IronPort-AV: E=Sophos;i="5.88,286,1635235200"; d="scan'208,223";a="73249740" Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa1.mentor.iphmx.com with ESMTP; 13 Jan 2022 04:22:28 -0800 IronPort-SDR: GmsLwL7lks4b393s8nZqZZhesyUkwt0sI0Wbuu5SaWxgufbtyAj0gS7MbkbNPLiBDjStVIlLa2 ZFHyFIYUx4AnfUoC/e4yTAO9fm4W8W7UrIzXE72drSvO2OCJOrTMocmuC98to9LzoSMWwf6/rr Dz+YsouC9UYl29hFbTNOW9LyOkDvAA5DWCQkBmj2Hmjo4Ztc1L/TmzToZCbqpTIysKCwgf5gcU g08i4KYwOZBf4pL2cASNJXNUblrYBmJNsrx21ZDIdLJOCqUSSt09tovUYmJcNsFfdV7Iw6y5s5 lc0= From: Thomas Schwinge To: Tobias Burnus , Chung-Lin Tang , CC: Jakub Jelinek , Andrew Stubbs , Marcel Vollweiler Subject: Re: [committed] libgomp/testsuite: Improve omp_get_device_num() tests In-Reply-To: <77fb4ef2-f2e8-5b79-aca4-085311d33a8b@codesourcery.com> References: <32c3546a-49c1-bedf-3002-c4e7783bf312@codesourcery.com> <268addd8-4c4f-6284-bf5c-df07e752811f@codesourcery.com> <77fb4ef2-f2e8-5b79-aca4-085311d33a8b@codesourcery.com> User-Agent: Notmuch/0.29.3+94~g74c3f1b (https://notmuchmail.org) Emacs/27.1 (x86_64-pc-linux-gnu) Date: Thu, 13 Jan 2022 13:22:18 +0100 Message-ID: <87r19bddvp.fsf@euler.schwinge.homeip.net> MIME-Version: 1.0 Content-Type: multipart/mixed; boundary="=-=-=" X-Originating-IP: [137.202.0.90] X-ClientProxiedBy: SVR-IES-MBX-08.mgc.mentorg.com (139.181.222.8) To svr-ies-mbx-01.mgc.mentorg.com (139.181.222.1) X-Spam-Status: No, score=-12.0 required=5.0 tests=BAYES_00, GIT_PATCH_0, HEADER_FROM_DIFFERENT_DOMAINS, KAM_DMARC_STATUS, KAM_LOTSOFHASH, SPF_HELO_PASS, SPF_PASS, TXREP autolearn=ham autolearn_force=no version=3.4.4 X-Spam-Checker-Version: SpamAssassin 3.4.4 (2020-01-24) on server2.sourceware.org X-BeenThere: gcc-patches@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Gcc-patches mailing list List-Unsubscribe: , List-Archive: List-Post: List-Help: List-Subscribe: , X-List-Received-Date: Thu, 13 Jan 2022 12:22:31 -0000 --=-=-= Content-Type: text/plain; charset="utf-8" Content-Transfer-Encoding: quoted-printable Hi! On 2022-01-04T15:12:58+0100, Tobias Burnus wrote: > This commit r12-6209 now makes the testcases iterate over all devices > (including the initial/host device). > > Hence, with multiple non-host devices and this test, the error had been > found before ... ;-) Yay for test cases! :-) ... but we now run into issues if Intel MIC (emulated) offloading is (additionally) enabled, because that one still doesn't properly implement device-side 'omp_get_device_num'. ;-) Thus pushed to master branch commit d97364aab1af361275b87713154c366ce2b9029a "Improve Intel MIC offloading XFAILing for 'omp_get_device_num'", see attached. (It wasn't obvious to me how to implement that; very incomplete "[WIP] Intel MIC 'omp_get_device_num'" attached, not planning on working on this any further.) Gr=C3=BC=C3=9Fe Thomas ----------------- Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstra=C3=9Fe 201= , 80634 M=C3=BCnchen; Gesellschaft mit beschr=C3=A4nkter Haftung; Gesch=C3= =A4ftsf=C3=BChrer: Thomas Heurung, Frank Th=C3=BCrauf; Sitz der Gesellschaf= t: M=C3=BCnchen; Registergericht M=C3=BCnchen, HRB 106955 --=-=-= Content-Type: text/x-diff Content-Disposition: inline; filename="0001-Improve-Intel-MIC-offloading-XFAILing-for-omp_get_de.patch" >From d97364aab1af361275b87713154c366ce2b9029a Mon Sep 17 00:00:00 2001 From: Thomas Schwinge Date: Wed, 5 Jan 2022 19:52:25 +0100 Subject: [PATCH] Improve Intel MIC offloading XFAILing for 'omp_get_device_num' After recent commit be661959a6b6d8f9c3c8608a746789e7b2ec3ca4 "libgomp/testsuite: Improve omp_get_device_num() tests", we're now iterating over all OpenMP target devices. Intel MIC (emulated) offloading still doesn't properly implement device-side 'omp_get_device_num', and we thus regress: PASS: libgomp.c/../libgomp.c-c++-common/target-45.c (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.c/../libgomp.c-c++-common/target-45.c execution test PASS: libgomp.c++/../libgomp.c-c++-common/target-45.c (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.c++/../libgomp.c-c++-common/target-45.c execution test PASS: libgomp.fortran/target10.f90 -O0 (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O0 execution test PASS: libgomp.fortran/target10.f90 -O1 (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O1 execution test PASS: libgomp.fortran/target10.f90 -O2 (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O2 execution test PASS: libgomp.fortran/target10.f90 -O3 -fomit-frame-pointer -funroll-loops -fpeel-loops -ftracer -finline-functions (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O3 -fomit-frame-pointer -funroll-loops -fpeel-loops -ftracer -finline-functions execution test PASS: libgomp.fortran/target10.f90 -O3 -g (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O3 -g execution test PASS: libgomp.fortran/target10.f90 -Os (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -Os execution test Improve the XFAILing added in commit bb75b22aba254e8ff144db27b1c8b4804bad73bb "Allow matching Intel MIC in OpenMP 'declare variant'" for the case that *any* Intel MIC offload device is available. libgomp/ * testsuite/libgomp.c-c++-common/on_device_arch.h (any_device_arch, any_device_arch_intel_mic): New. * testsuite/lib/libgomp.exp (check_effective_target_offload_device_any_intel_mic): New. * testsuite/libgomp.c-c++-common/target-45.c: Use it. * testsuite/libgomp.fortran/target10.f90: Likewise. --- libgomp/testsuite/lib/libgomp.exp | 12 +++++++++- .../libgomp.c-c++-common/on_device_arch.h | 23 +++++++++++++++++++ .../libgomp.c-c++-common/target-45.c | 2 +- .../testsuite/libgomp.fortran/target10.f90 | 2 +- 4 files changed, 36 insertions(+), 3 deletions(-) diff --git a/libgomp/testsuite/lib/libgomp.exp b/libgomp/testsuite/lib/libgomp.exp index 57fb6b068f3..8c5ecfff0ac 100644 --- a/libgomp/testsuite/lib/libgomp.exp +++ b/libgomp/testsuite/lib/libgomp.exp @@ -451,7 +451,6 @@ proc check_effective_target_openacc_nvidia_accel_selected { } { # Return 1 if using Intel MIC offload device. proc check_effective_target_offload_device_intel_mic { } { return [check_runtime_nocache offload_device_intel_mic { - #include #include "testsuite/libgomp.c-c++-common/on_device_arch.h" int main () { @@ -460,6 +459,17 @@ proc check_effective_target_offload_device_intel_mic { } { } ] } +# Return 1 if any Intel MIC offload device is available. +proc check_effective_target_offload_device_any_intel_mic { } { + return [check_runtime_nocache offload_device_any_intel_mic { + #include "testsuite/libgomp.c-c++-common/on_device_arch.h" + int main () + { + return !any_device_arch_intel_mic (); + } + } ] +} + # Return 1 if the OpenACC 'host' device type is selected. proc check_effective_target_openacc_host_selected { } { diff --git a/libgomp/testsuite/libgomp.c-c++-common/on_device_arch.h b/libgomp/testsuite/libgomp.c-c++-common/on_device_arch.h index ee541dd2260..f92743b04d7 100644 --- a/libgomp/testsuite/libgomp.c-c++-common/on_device_arch.h +++ b/libgomp/testsuite/libgomp.c-c++-common/on_device_arch.h @@ -1,3 +1,4 @@ +#include #include /* static */ int @@ -41,3 +42,25 @@ on_device_arch_intel_mic () { return on_device_arch (GOMP_DEVICE_INTEL_MIC); } + +static int +any_device_arch (int d) +{ + int nd = omp_get_num_devices (); + for (int i = 0; i < nd; ++i) + { + int d_cur; + #pragma omp target device(i) map(from:d_cur) + d_cur = device_arch (); + if (d_cur == d) + return 1; + } + + return 0; +} + +int +any_device_arch_intel_mic () +{ + return any_device_arch (GOMP_DEVICE_INTEL_MIC); +} diff --git a/libgomp/testsuite/libgomp.c-c++-common/target-45.c b/libgomp/testsuite/libgomp.c-c++-common/target-45.c index 837503996d7..27bbeddf7fd 100644 --- a/libgomp/testsuite/libgomp.c-c++-common/target-45.c +++ b/libgomp/testsuite/libgomp.c-c++-common/target-45.c @@ -1,4 +1,4 @@ -/* { dg-xfail-run-if TODO { offload_device_intel_mic } } */ +/* { dg-xfail-run-if TODO { offload_device_any_intel_mic } } */ #include #include diff --git a/libgomp/testsuite/libgomp.fortran/target10.f90 b/libgomp/testsuite/libgomp.fortran/target10.f90 index f6951fc9057..31452554d67 100644 --- a/libgomp/testsuite/libgomp.fortran/target10.f90 +++ b/libgomp/testsuite/libgomp.fortran/target10.f90 @@ -1,5 +1,5 @@ ! { dg-do run } -! { dg-xfail-run-if TODO { offload_device_intel_mic } } +! { dg-xfail-run-if TODO { offload_device_any_intel_mic } } program main use omp_lib -- 2.34.1 --=-=-= Content-Type: text/x-diff Content-Disposition: inline; filename="0001-WIP-Intel-MIC-omp_get_device_num.patch" >From cea0fd2d10f7a9fd060543e59e142c8c9d06abb0 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge Date: Wed, 5 Jan 2022 19:15:06 +0100 Subject: [PATCH] [WIP] Intel MIC 'omp_get_device_num' See commit 0bac793ed6bad2c0c13cd1e93a1aa5808467afc8 "openmp: Implement omp_get_device_num routine" P commit fbb592407c9dd244b4cea086cbb90d7bd0bf60bb "libgomp: Fix GOMP_DEVICE_NUM_VAR stringification during offload image load" P commit be661959a6b6d8f9c3c8608a746789e7b2ec3ca4 "libgomp/testsuite: Improve omp_get_device_num() tests" With Intel MIC (emulated) offloading enabled: PASS: libgomp.c/../libgomp.c-c++-common/target-45.c (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.c/../libgomp.c-c++-common/target-45.c execution test PASS: libgomp.c++/../libgomp.c-c++-common/target-45.c (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.c++/../libgomp.c-c++-common/target-45.c execution test PASS: libgomp.fortran/target10.f90 -O0 (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O0 execution test PASS: libgomp.fortran/target10.f90 -O1 (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O1 execution test PASS: libgomp.fortran/target10.f90 -O2 (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O2 execution test PASS: libgomp.fortran/target10.f90 -O3 -fomit-frame-pointer -funroll-loops -fpeel-loops -ftracer -finline-functions (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O3 -fomit-frame-pointer -funroll-loops -fpeel-loops -ftracer -finline-functions execution test PASS: libgomp.fortran/target10.f90 -O3 -g (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -O3 -g execution test PASS: libgomp.fortran/target10.f90 -Os (test for excess errors) [-PASS:-]{+FAIL:+} libgomp.fortran/target10.f90 -Os execution test TODO Not sure if this is the right approach for Intel MIC, over all. TODO This is, at least, still missing corresponding changes to 'liboffloadmic/plugin/libgomp-plugin-intelmic.cpp': 'GOMP_OFFLOAD_load_image'/'offload_image'. TODO I don't understand how 'libgomp/icv-device.c'/'libgomp/config/*/icv-device.c' are split over 'liboffloadmic/plugin/offload_target_main.cpp' vs. 'liboffloadmic/runtime/offload_omp_host.cpp'/'liboffloadmic/runtime/offload_omp_target.cpp'. --- liboffloadmic/plugin/Makefile.am | 2 +- liboffloadmic/plugin/Makefile.in | 4 +++- liboffloadmic/plugin/offload_target_main.cpp | 24 +++++++++++++++++++- 3 files changed, 27 insertions(+), 3 deletions(-) diff --git a/liboffloadmic/plugin/Makefile.am b/liboffloadmic/plugin/Makefile.am index 7caea7894ac..840940cb760 100644 --- a/liboffloadmic/plugin/Makefile.am +++ b/liboffloadmic/plugin/Makefile.am @@ -57,7 +57,7 @@ if PLUGIN_HOST else # PLUGIN_TARGET plugin_includedir = $(libsubincludedir) plugin_include_HEADERS = main_target_image.h - AM_CPPFLAGS = $(CPPFLAGS) -DLINUX -DCOI_LIBRARY_VERSION=2 -DOFFLOAD_DEBUG=1 -DSEP_SUPPORT -DTIMING_SUPPORT -DHOST_LIBRARY=0 -I$(coi_inc_dir) -I$(liboffload_src_dir) -I$(libgomp_dir) + AM_CPPFLAGS = $(CPPFLAGS) -DLINUX -DCOI_LIBRARY_VERSION=2 -DOFFLOAD_DEBUG=1 -DSEP_SUPPORT -DTIMING_SUPPORT -DHOST_LIBRARY=0 -I$(coi_inc_dir) -I$(liboffload_src_dir) -I$(libgomp_src_dir) -I$(libgomp_dir) AM_CXXFLAGS = $(CXXFLAGS) AM_LDFLAGS = -L$(liboffload_dir)/.libs -L$(libgomp_dir)/.libs -loffloadmic_target -lcoi_device -lgomp -rdynamic endif diff --git a/liboffloadmic/plugin/Makefile.in b/liboffloadmic/plugin/Makefile.in index 8d5ad0025c2..45dcd01bab3 100644 --- a/liboffloadmic/plugin/Makefile.in +++ b/liboffloadmic/plugin/Makefile.in @@ -401,7 +401,7 @@ target_install_dir = $(accel_search_dir)/lib/gcc/$(accel_target)/$(gcc_version)$ @PLUGIN_HOST_TRUE@libgomp_plugin_intelmic_la_LDFLAGS = -L$(liboffload_dir)/.libs -loffloadmic_host -version-info 1:0:0 @PLUGIN_HOST_FALSE@plugin_includedir = $(libsubincludedir) @PLUGIN_HOST_FALSE@plugin_include_HEADERS = main_target_image.h -@PLUGIN_HOST_FALSE@AM_CPPFLAGS = $(CPPFLAGS) -DLINUX -DCOI_LIBRARY_VERSION=2 -DOFFLOAD_DEBUG=1 -DSEP_SUPPORT -DTIMING_SUPPORT -DHOST_LIBRARY=0 -I$(coi_inc_dir) -I$(liboffload_src_dir) -I$(libgomp_dir) +@PLUGIN_HOST_FALSE@AM_CPPFLAGS = $(CPPFLAGS) -DLINUX -DCOI_LIBRARY_VERSION=2 -DOFFLOAD_DEBUG=1 -DSEP_SUPPORT -DTIMING_SUPPORT -DHOST_LIBRARY=0 -I$(coi_inc_dir) -I$(liboffload_src_dir) -I$(libgomp_src_dir) -I$(libgomp_dir) @PLUGIN_HOST_FALSE@AM_CXXFLAGS = $(CXXFLAGS) @PLUGIN_HOST_FALSE@AM_LDFLAGS = -L$(liboffload_dir)/.libs -L$(libgomp_dir)/.libs -loffloadmic_target -lcoi_device -lgomp -rdynamic @@ -838,3 +838,5 @@ maintainer-clean-local: maintainer-clean-multi # Tell versions [3.59,3.63) of GNU make to not export all variables. # Otherwise a system limit (for SysV at least) may be exceeded. .NOEXPORT: + +#TODO Properly regenerate. diff --git a/liboffloadmic/plugin/offload_target_main.cpp b/liboffloadmic/plugin/offload_target_main.cpp index ccf7240e286..3b7e06e249c 100644 --- a/liboffloadmic/plugin/offload_target_main.cpp +++ b/liboffloadmic/plugin/offload_target_main.cpp @@ -31,6 +31,7 @@ #include #include #include +#include "libgomp-plugin.h" #include "compiler_if_target.h" @@ -91,7 +92,10 @@ static void *last_var_ptr = NULL; static int last_var_size = 0; -/* Override the corresponding functions from libgomp. */ +/* Override functions from 'libgomp/icv-device.c' (see + 'libgomp/config/[...]/icv-device.c') as well as the corresponding + 'libgomp/fortran.c' wrapper routines. */ + extern "C" int omp_is_initial_device (void) __GOMP_NOTHROW { @@ -104,6 +108,24 @@ omp_is_initial_device_ (void) return omp_is_initial_device (); } +#ifndef GOMP_DEVICE_NUM_VAR +# error +#endif + +static volatile int GOMP_DEVICE_NUM_VAR; + +extern "C" int +omp_get_device_num (void) __GOMP_NOTHROW +{ + return GOMP_DEVICE_NUM_VAR; +} + +extern "C" int32_t +omp_get_device_num_ (void) +{ + return omp_get_device_num (); +} + /* Dummy function needed for the initialization of target process during the first call to __offload_offload1. */ -- 2.34.1 --=-=-=--