From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: from esa3.mentor.iphmx.com (esa3.mentor.iphmx.com [68.232.137.180]) by sourceware.org (Postfix) with ESMTPS id 68B7D381EC8C; Wed, 7 Dec 2022 16:13:53 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.1 sourceware.org 68B7D381EC8C Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=mentor.com X-IronPort-AV: E=Sophos;i="5.96,225,1665475200"; d="scan'208";a="88702727" Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa3.mentor.iphmx.com with ESMTP; 07 Dec 2022 08:13:50 -0800 IronPort-SDR: 9Wja3rd7UWsf0KvkldGggH+qB7aBrSbdgtWVlWbp7PSQjvDoEaJicsBwYZ/6M4OcbjteQQBmKQ 178QkrqoiE4S3H4x8zaaZa2omnoHatVGaN+mbwTOj1jnBBhuqtyvLmalveQlwYa/Msf5wM8Ouv ChrzMubWqr6azrwIyusURR6siwPOste79C7lCHwbFJKStDD19alsmTCxgrFZhGcLZM+K/gk4ij ruuWmhNjQuqEkK/BRcaDHJFlfeq+NiLXasCv+I2kIFo3AxfEJ8C8HYYHJdzfmPGA/qLhnrcaaf I7o= Message-ID: <81d1d3c5-4f37-21ca-f6e5-bd75bc197c2b@codesourcery.com> Date: Wed, 7 Dec 2022 17:13:42 +0100 MIME-Version: 1.0 User-Agent: Mozilla/5.0 (X11; Linux x86_64; rv:102.0) Gecko/20100101 Thunderbird/102.5.1 Subject: Re: [PATCH v5 2/4] OpenMP/OpenACC: Rework clause expansion and nested struct handling Content-Language: en-US To: Julian Brown CC: , , Jakub Jelinek , Chung-Lin Tang References: <8fcf3df1b40ea77cbb8088962cbcdf6935d2ded3.1666088224.git.julian@codesourcery.com> <65824aa6-be83-3c74-871e-6571008f2d25@codesourcery.com> <20221207151657.7340469c@squid.athome> From: Tobias Burnus In-Reply-To: <20221207151657.7340469c@squid.athome> Content-Type: text/plain; charset="UTF-8"; format=flowed Content-Transfer-Encoding: quoted-printable X-Originating-IP: [137.202.0.90] X-ClientProxiedBy: SVR-IES-MBX-07.mgc.mentorg.com (139.181.222.7) To svr-ies-mbx-12.mgc.mentorg.com (139.181.222.12) X-Spam-Status: No, score=-5.5 required=5.0 tests=BAYES_00,HEADER_FROM_DIFFERENT_DOMAINS,KAM_DMARC_STATUS,KAM_SHORT,NICE_REPLY_A,SPF_HELO_PASS,SPF_PASS,TXREP autolearn=ham autolearn_force=no version=3.4.6 X-Spam-Checker-Version: SpamAssassin 3.4.6 (2021-04-09) on server2.sourceware.org List-Id: Hi Julian, On 07.12.22 16:16, Julian Brown wrote: > On Wed, 7 Dec 2022 15:54:42 +0100 Tobias Burnus = wrote: >> If I understand Deepak's comment (on OpenMP.org's omp-lang list, sorry >> it is a nonpublic list) correctly, the following wording implies that >> a 'from: s.w[z:4]' for a pointer 's.w' also implies a mapping of >> 's.w' - if 's' is used inside the target region and, thus, gets >> implicitly mapped. >> >> [TR11 157:21-26] (approx. [5.2 154:22-27], [5.1 352:17-22], [5.0 >> 320:22-27]) >> >> "If a list item with an implicit data-mapping attribute does not have >> any corresponding storage in the device data environment prior to a >> task encountering the construct associated with the map clause, and >> one or more contiguous parts of the original storage are either list >> items or base pointers to list items that are explicitly mapped on >> the construct, only those parts of the original storage will have >> corresponding storage in the device data environment as a result of >> the map clauses on the construct." > Hmmm... IIRC that is a different conclusion than the one we have > understood previously, leading to e.g. the patch here (Chung-Lin CC'ed): > > https://gcc.gnu.org/pipermail/gcc-patches/2021-May/570075.html This seems to be the "Target directive struct mapping question" omp-lang th= read, started on 2021-03-22. I think we need to distinguish: #pragma omp target enter data map(to: s.w[:10]) from #pragma omp target map(tofrom: s.arr[:20]) s.arr[0] =3D 5; As in the latter case 's' gets implicitly mapped and then applies to the base pointer 's.arr' of 's.arr[:20]'. While in the former case, only the pointee gets mapped without the pointer 's.arr' (and, hence, there is also no pointer attachment). At least that's what I get from the wording above and reading Deepak's last email - and it does not seem to clash with the discussion in the lengthy omp-lang thread. (Maybe there are other threads =E2=80=93 or I completely m= isread them.) I think it makes sense to have a clarifying example in OpenMP; hence, I filed the OpenMP.org example issue #342, starting with essentially what I wrote above: 'target enter data' needs more work to get the pointer handling done, 'target' + accessing 's' works as is. I hope it makes sense. > Follow-on discussion then questioned whether the change was really the > intention of the spec, but we thought it was. Has that changed now? No idea =E2=80=93 I find it difficult to track all the language changes and= find mapping complex and unclear. However, it does seem to make sense in the way written above without contradicting to all previous discussions, minus the common confusion. (As least as I gathered from browsing both omp-lang and gcc-patches.) > (I think actually changing the behaviour is a matter of flipping a > switch, but let's make sure we choose the right setting!) That sounds great! Tobias ----------------- 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