From patchwork Tue Oct 13 15:49:21 2015 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: Tom de Vries X-Patchwork-Id: 529839 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Received: from sourceware.org (server1.sourceware.org [209.132.180.131]) (using TLSv1.2 with cipher ECDHE-RSA-AES256-GCM-SHA384 (256/256 bits)) (No client certificate requested) by ozlabs.org (Postfix) with ESMTPS id 0AB961402C4 for ; Wed, 14 Oct 2015 02:50:33 +1100 (AEDT) Authentication-Results: ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=gcc.gnu.org header.i=@gcc.gnu.org header.b=e4hIcRFs; dkim-atps=neutral DomainKey-Signature: a=rsa-sha1; c=nofws; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender :subject:to:references:cc:from:message-id:date:mime-version :in-reply-to:content-type; q=dns; s=default; b=rQ35kJlWmG5oDPADv jngrhtUVBfEYxCOfnTJXyAgP0y0iJp0t1m0B1iDlUeA751I09jrjPvXdmiCWjCiH xckjnrutiicyHToszcQeMiRW2RW226XBm7Hj3W0xHlGHD0/TBFONEwmGRuGF0Qd3 C8Ht0n1XWyGodp3SXhVOJ32FAs= DKIM-Signature: v=1; a=rsa-sha1; c=relaxed; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender :subject:to:references:cc:from:message-id:date:mime-version :in-reply-to:content-type; s=default; bh=VX6YBf7+aSXeHfU847O0zMx bbTg=; b=e4hIcRFsLmfV8eUVQp/feR4XlrUs+2vYw+ndzbRlnDsfhX1KoPWZ4HW sOhsfeev38R99i0w7EvqMPO/21F3D8GKNTKuCfjS30GrKQrUVdo/buTVVWHNSYRN yNPObWP1rgcvOTbZd/5Lj1eeLHoK+E1SgoV3NXifr8Pypi/hMZZ0= Received: (qmail 94796 invoked by alias); 13 Oct 2015 15:50:27 -0000 Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Unsubscribe: List-Archive: List-Post: List-Help: Sender: gcc-patches-owner@gcc.gnu.org Delivered-To: mailing list gcc-patches@gcc.gnu.org Received: (qmail 94786 invoked by uid 89); 13 Oct 2015 15:50:26 -0000 Authentication-Results: sourceware.org; auth=none X-Virus-Found: No X-Spam-SWARE-Status: No, score=-2.0 required=5.0 tests=AWL, BAYES_00, SPF_PASS, T_RP_MATCHES_RCVD autolearn=ham version=3.3.2 X-HELO: fencepost.gnu.org Received: from fencepost.gnu.org (HELO fencepost.gnu.org) (208.118.235.10) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with (AES128-SHA encrypted) ESMTPS; Tue, 13 Oct 2015 15:50:24 +0000 Received: from eggs.gnu.org ([2001:4830:134:3::10]:59813) by fencepost.gnu.org with esmtps (TLS1.0:RSA_AES_256_CBC_SHA1:256) (Exim 4.82) (envelope-from ) id 1Zm1qI-0000vS-Fq for gcc-patches@gnu.org; Tue, 13 Oct 2015 11:50:22 -0400 Received: from Debian-exim by eggs.gnu.org with spam-scanned (Exim 4.71) (envelope-from ) id 1Zm1qC-000526-Py for gcc-patches@gnu.org; Tue, 13 Oct 2015 11:50:22 -0400 Received: from relay1.mentorg.com ([192.94.38.131]:45006) by eggs.gnu.org with esmtp (Exim 4.71) (envelope-from ) id 1Zm1qC-00051M-HT for gcc-patches@gnu.org; Tue, 13 Oct 2015 11:50:16 -0400 Received: from nat-ies.mentorg.com ([192.94.31.2] helo=SVR-IES-FEM-01.mgc.mentorg.com) by relay1.mentorg.com with esmtp id 1Zm1q6-00033g-3n from Tom_deVries@mentor.com for gcc-patches@gnu.org; Tue, 13 Oct 2015 08:50:10 -0700 Received: from [127.0.0.1] (137.202.0.76) by SVR-IES-FEM-01.mgc.mentorg.com (137.202.0.104) with Microsoft SMTP Server id 14.3.224.2; Tue, 13 Oct 2015 16:50:08 +0100 Subject: Re: [gomp4, committed] Add goacc/kernels-acc-on-device.c To: Thomas Schwinge References: <5618ED1D.60906@mentor.com> <874mhwmkf0.fsf@kepler.schwinge.homeip.net> <561BACFD.6030001@mentor.com> CC: "gcc-patches@gnu.org" From: Tom de Vries Message-ID: <561D2801.1070004@mentor.com> Date: Tue, 13 Oct 2015 17:49:21 +0200 User-Agent: Mozilla/5.0 (X11; Linux x86_64; rv:38.0) Gecko/20100101 Thunderbird/38.3.0 MIME-Version: 1.0 In-Reply-To: <561BACFD.6030001@mentor.com> X-detected-operating-system: by eggs.gnu.org: Windows NT kernel [generic] [fuzzy] X-Received-From: 192.94.38.131 On 12/10/15 14:52, Tom de Vries wrote: > On 12/10/15 12:49, Thomas Schwinge wrote: >> Hi Tom! >> >> On Sat, 10 Oct 2015 12:49:01 +0200, Tom de >> Vries wrote: >>> >--- /dev/null >>> >+++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c >>> >@@ -0,0 +1,39 @@ >>> >+/* { dg-additional-options "-O2" } */ >>> >+ >>> >+#include > > Hi Thomas, > >> That doesn't work (at least in build-tree testing), as gcc/testsuite/ is >> not set up to look for header files in [target]/libgomp/: >> >> [...]/source-gcc/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c:3:21: >> fatal error: openacc.h: No such file or directory >> compilation terminated. >> compiler exited with status 1 >> > > Ah, I see. I was doing 'make' followed by 'make install', and then > build-tree testing. The build-tree testing seems to pick up the header > file from the install directory. So for me test passed. > >>> >+ >>> >+#define N 32 >>> >+ >>> >+void >>> >+foo (float *a, float *b) >>> >+{ >>> >+ float exp; >>> >+ int i; >>> >+ int n; >>> >+ >>> >+#pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) >>> >+ { >>> >+ int ii; >>> >+ >>> >+ for (ii = 0; ii < N; ii++) >>> >+ { >>> >+ if (acc_on_device (acc_device_host)) >> Your two options are: if that's applicable/sufficient for what you intend >> to test here, use __builtin_acc_on_device with a hard-coded acc_device_*, >> or duplicate part of as done for example in >> gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c. >> > > Went with second option, committed as attached. As a follow-up patch, I've factored the code into a mockup openacc.h, now shared by several test-cases. Committed to gomp-4_0-branch. Thanks, - Tom Factor out goacc/openacc.h 2015-10-13 Tom de Vries * c-c++-common/goacc/openacc.h: New header file, factored out of ... * c-c++-common/goacc/kernels-acc-on-device.c: ... here. * c-c++-common/goacc/acc_on_device-2-off.c: Use openacc.h. * c-c++-common/goacc/acc_on_device-2.c: Same. --- .../c-c++-common/goacc/acc_on_device-2-off.c | 11 +---------- gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c | 13 +------------ .../c-c++-common/goacc/kernels-acc-on-device.c | 19 +------------------ gcc/testsuite/c-c++-common/goacc/openacc.h | 18 ++++++++++++++++++ 4 files changed, 21 insertions(+), 40 deletions(-) create mode 100644 gcc/testsuite/c-c++-common/goacc/openacc.h diff --git a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c index 71abe11..cce58de 100644 --- a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c +++ b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c @@ -3,16 +3,7 @@ /* Duplicate parts of libgomp/openacc.h, because we can't include it here. */ -#if __cplusplus -extern "C" { -#endif - -typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; -extern int acc_on_device (int); - -#if __cplusplus -} -#endif +#include "openacc.h" int f (void) diff --git a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c index 243e562..19a5bd3 100644 --- a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c +++ b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c @@ -1,18 +1,7 @@ /* Have to enable optimizations, as otherwise builtins won't be expanded. */ /* { dg-additional-options "-O -fdump-rtl-expand" } */ -/* Duplicate parts of libgomp/openacc.h, because we can't include it here. */ - -#if __cplusplus -extern "C" { -#endif - -typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; -extern int acc_on_device (int); - -#if __cplusplus -} -#endif +#include "openacc.h" int f (void) diff --git a/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c index 784c66a..958b65b 100644 --- a/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c +++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c @@ -1,23 +1,6 @@ /* { dg-additional-options "-O2" } */ -#if __cplusplus -extern "C" { -#endif - -#if __cplusplus >= 201103 -# define __GOACC_NOTHROW noexcept -#elif __cplusplus -# define __GOACC_NOTHROW throw () -#else /* Not C++ */ -# define __GOACC_NOTHROW __attribute__ ((__nothrow__)) -#endif - -typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; -int acc_on_device (int) __GOACC_NOTHROW; - -#if __cplusplus -} -#endif +#include "openacc.h" #define N 32 diff --git a/gcc/testsuite/c-c++-common/goacc/openacc.h b/gcc/testsuite/c-c++-common/goacc/openacc.h new file mode 100644 index 0000000..a74a482 --- /dev/null +++ b/gcc/testsuite/c-c++-common/goacc/openacc.h @@ -0,0 +1,18 @@ +#if __cplusplus +extern "C" { +#endif + +#if __cplusplus >= 201103 +# define __GOACC_NOTHROW noexcept +#elif __cplusplus +# define __GOACC_NOTHROW throw () +#else /* Not C++ */ +# define __GOACC_NOTHROW __attribute__ ((__nothrow__)) +#endif + +typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; +int acc_on_device (int) __GOACC_NOTHROW; + +#if __cplusplus +} +#endif -- 1.9.1