From patchwork Mon May 10 14:11:39 2021 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Marcel Vollweiler X-Patchwork-Id: 1476459 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Authentication-Results: ozlabs.org; spf=pass (sender SPF authorized) smtp.mailfrom=gcc.gnu.org (client-ip=8.43.85.97; helo=sourceware.org; envelope-from=gcc-patches-bounces@gcc.gnu.org; receiver=) Received: from sourceware.org (server2.sourceware.org [8.43.85.97]) (using TLSv1.3 with cipher TLS_AES_256_GCM_SHA384 (256/256 bits) key-exchange X25519 server-signature RSA-PSS (4096 bits) server-digest SHA256) (No client certificate requested) by ozlabs.org (Postfix) with ESMTPS id 4Ff2zr4t0Yz9sWk for ; Tue, 11 May 2021 00:11:51 +1000 (AEST) Received: from server2.sourceware.org (localhost [IPv6:::1]) by sourceware.org (Postfix) with ESMTP id 7CB5A3851C13; Mon, 10 May 2021 14:11:49 +0000 (GMT) X-Original-To: gcc-patches@gcc.gnu.org Delivered-To: gcc-patches@gcc.gnu.org Received: from esa3.mentor.iphmx.com (esa3.mentor.iphmx.com [68.232.137.180]) by sourceware.org (Postfix) with ESMTPS id A99373857813 for ; Mon, 10 May 2021 14:11:45 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.3.2 sourceware.org A99373857813 Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=Marcel_Vollweiler@mentor.com IronPort-SDR: v+/BUujbGpQxf7Sl5bdkgh9ktcecYz0Y2kugUXCWK3CDhBmoa1M+xXCUqHiZwt5YTKET32/iIH mRCBQ3qnsN6EVnPSk2qStLbh2zWIsmGUVoVT63CX2ILtvbqTqeDzd46Spcs5sWEnbye35QzmaV z07gWbcGvhY5jQeC6iBf8n9jBND5Kiul70wHudLRaSWYsm/ZzO+gj8EJOsemuSMFeZ/hYBUlLI 6ohkQT1tbaRDPRLKL8JooJvLmhsM1/3Yhfj9PuetNAnPehcQDUhblDtCdALM/9vGuyW98+3YTP Jw0= X-IronPort-AV: E=Sophos;i="5.82,287,1613462400"; d="diff'?scan'208";a="61031243" Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa3.mentor.iphmx.com with ESMTP; 10 May 2021 06:11:44 -0800 IronPort-SDR: 6UYFQiA5Z0Sqn19G/fCwIZw4yeq+S+5+uvJxo7zxlTX1kY0Aj9EAg8+ERr6Lk5UiG8k14wOI// G8WM/Mm5gL7c0P/mhohqOXeTi/O++PGX2/4d7Vz14lwNKxc+8y0zCphBJOMBnPw0dRIhQ+tr2J pwXuX5ThJXXsfEKT5Lhh79ZVenv13g5Nts0zqLK77+rftSq7IQe8BVHlHQZv2pba4SaFfPHuUx OfntQ5hEAX6SlIi4sbfvjP/q2OKCPCp71SCyiJNFzkYhAWl/mMpU2OOJkQuQoFgPqQZo8aOmHJ 5zk= To: , From: Marcel Vollweiler Subject: [PATCH] OpenMP: Add support for 'close' in map clause Message-ID: Date: Mon, 10 May 2021 16:11:39 +0200 User-Agent: Mozilla/5.0 (Windows NT 10.0; Win64; x64; rv:78.0) Gecko/20100101 Thunderbird/78.10.0 MIME-Version: 1.0 X-Originating-IP: [137.202.0.90] X-ClientProxiedBy: svr-ies-mbx-05.mgc.mentorg.com (139.181.222.5) To SVR-IES-MBX-03.mgc.mentorg.com (139.181.222.3) X-Spam-Status: No, score=-13.4 required=5.0 tests=BAYES_00, GIT_PATCH_0, HEADER_FROM_DIFFERENT_DOMAINS, KAM_DMARC_STATUS, SPF_HELO_PASS, SPF_PASS, TXREP autolearn=ham autolearn_force=no version=3.4.2 X-Spam-Checker-Version: SpamAssassin 3.4.2 (2018-09-13) 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: , Errors-To: gcc-patches-bounces@gcc.gnu.org Sender: "Gcc-patches" Hi, This patch adds handling for the map-type-modifier 'close' in the map clause that was introduced with OpenMP 5.0: "The close map-type-modifier is a hint to the runtime to allocate memory close to the target device." In OpenMP 5.0 'close' can be used beside/together with 'always' in a list of map-type-modifiers. With this patch, 'close' will be parsed and ignored for C and C++. A patch for Fortran will be provided separately. Marcel ----------------- Mentor Graphics (Deutschland) GmbH, Arnulfstrasse 201, 80634 München Registergericht München HRB 106955, Geschäftsführer: Thomas Heurung, Frank Thürauf OpenMP: Add support for 'close' in map clause gcc/c/ChangeLog: * c-parser.c (c_parser_omp_clause_map): Support map-type-modifier 'close'. gcc/cp/ChangeLog: * parser.c (cp_parser_omp_clause_map): Support map-type-modifier 'close'. gcc/testsuite/ChangeLog: * c-c++-common/gomp/map-6.c: New test. * c-c++-common/gomp/map-7.c: New test. diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c index 5cdeb21..78cba7f 100644 --- a/gcc/c/c-parser.c +++ b/gcc/c/c-parser.c @@ -15643,14 +15643,19 @@ c_parser_omp_clause_depend (c_parser *parser, tree list) map-kind: alloc | to | from | tofrom | release | delete - map ( always [,] map-kind: variable-list ) */ + map ( always [,] map-kind: variable-list ) + + OpenMP 5.0: + map ( [map-type-modifier[,] ...] map-kind: variable-list ) + + map-type-modifier: + always | close */ static tree c_parser_omp_clause_map (c_parser *parser, tree list) { location_t clause_loc = c_parser_peek_token (parser)->location; enum gomp_map_kind kind = GOMP_MAP_TOFROM; - int always = 0; enum c_id_kind always_id_kind = C_ID_NONE; location_t always_loc = UNKNOWN_LOCATION; tree always_id = NULL_TREE; @@ -15660,37 +15665,54 @@ c_parser_omp_clause_map (c_parser *parser, tree list) if (!parens.require_open (parser)) return list; - if (c_parser_next_token_is (parser, CPP_NAME)) + int always = 0; + int close = 0; + int pos = 1; + while (c_parser_peek_nth_token_raw (parser, pos)->type == CPP_NAME) { - c_token *tok = c_parser_peek_token (parser); + c_token *tok = c_parser_peek_nth_token_raw (parser, pos); const char *p = IDENTIFIER_POINTER (tok->value); - always_id_kind = tok->id_kind; - always_loc = tok->location; - always_id = tok->value; if (strcmp ("always", p) == 0) { - c_token *sectok = c_parser_peek_2nd_token (parser); - if (sectok->type == CPP_COMMA) + if (always) { - c_parser_consume_token (parser); - c_parser_consume_token (parser); - always = 2; + c_parser_error (parser, "expected modifier % only once"); + parens.skip_until_found_close (parser); + return list; + } + + always_id_kind = tok->id_kind; + always_loc = tok->location; + always_id = tok->value; + + always++; + } + else if (strcmp ("close", p) == 0) + { + if (close) + { + c_parser_error (parser, "expected modifier % only once"); + parens.skip_until_found_close (parser); + return list; } - else if (sectok->type == CPP_NAME) + + close++; + } + else if (c_parser_peek_nth_token_raw (parser, pos + 1)->type == CPP_COLON) + { + for (int i = 1; i < pos; ++i) { - p = IDENTIFIER_POINTER (sectok->value); - if (strcmp ("alloc", p) == 0 - || strcmp ("to", p) == 0 - || strcmp ("from", p) == 0 - || strcmp ("tofrom", p) == 0 - || strcmp ("release", p) == 0 - || strcmp ("delete", p) == 0) - { - c_parser_consume_token (parser); - always = 1; - } + c_parser_peek_token(parser); + c_parser_consume_token (parser); } + break; } + else + break; + + if (c_parser_peek_nth_token_raw (parser, pos + 1)->type == CPP_COMMA) + pos++; + pos++; } if (c_parser_next_token_is (parser, CPP_NAME) @@ -15719,35 +15741,6 @@ c_parser_omp_clause_map (c_parser *parser, tree list) c_parser_consume_token (parser); c_parser_consume_token (parser); } - else if (always) - { - if (always_id_kind != C_ID_ID) - { - c_parser_error (parser, "expected identifier"); - parens.skip_until_found_close (parser); - return list; - } - - tree t = lookup_name (always_id); - if (t == NULL_TREE) - { - undeclared_variable (always_loc, always_id); - t = error_mark_node; - } - if (t != error_mark_node) - { - tree u = build_omp_clause (clause_loc, OMP_CLAUSE_MAP); - OMP_CLAUSE_DECL (u) = t; - OMP_CLAUSE_CHAIN (u) = list; - OMP_CLAUSE_SET_MAP_KIND (u, kind); - list = u; - } - if (always == 1) - { - parens.skip_until_found_close (parser); - return list; - } - } nl = c_parser_omp_variable_list (parser, clause_loc, OMP_CLAUSE_MAP, list); diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c index 99eccf0..f7fecf7 100644 --- a/gcc/cp/parser.c +++ b/gcc/cp/parser.c @@ -37840,7 +37840,13 @@ cp_parser_omp_clause_depend (cp_parser *parser, tree list, location_t loc) map-kind: alloc | to | from | tofrom | release | delete - map ( always [,] map-kind: variable-list ) */ + map ( always [,] map-kind: variable-list ) + + OpenMP 5.0: + map ( [map-type-modifier[,] ...] map-kind: variable-list ) + + map-type-modifier: + always | close */ static tree cp_parser_omp_clause_map (cp_parser *parser, tree list) @@ -37848,32 +37854,61 @@ cp_parser_omp_clause_map (cp_parser *parser, tree list) tree nlist, c; enum gomp_map_kind kind = GOMP_MAP_TOFROM; bool always = false; + bool close = false; + int pos = 1; if (!cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN)) return list; - if (cp_lexer_next_token_is (parser->lexer, CPP_NAME)) + while (cp_lexer_peek_nth_token (parser->lexer, pos)->type == CPP_NAME + || cp_lexer_peek_nth_token (parser->lexer, pos)->keyword == RID_DELETE) { - tree id = cp_lexer_peek_token (parser->lexer)->u.value; + tree id = cp_lexer_peek_nth_token (parser->lexer, pos)->u.value; const char *p = IDENTIFIER_POINTER (id); if (strcmp ("always", p) == 0) { - int nth = 2; - if (cp_lexer_peek_nth_token (parser->lexer, 2)->type == CPP_COMMA) - nth++; - if ((cp_lexer_peek_nth_token (parser->lexer, nth)->type == CPP_NAME - || (cp_lexer_peek_nth_token (parser->lexer, nth)->keyword - == RID_DELETE)) - && (cp_lexer_peek_nth_token (parser->lexer, nth + 1)->type - == CPP_COLON)) + if (always) { - always = true; - cp_lexer_consume_token (parser->lexer); - if (nth == 3) - cp_lexer_consume_token (parser->lexer); + cp_parser_error (parser, + "expected modifier % only once"); + cp_parser_skip_to_closing_parenthesis (parser, + /*recovering=*/true, + /*or_comma=*/false, + /*consume_paren=*/true); + return list; + } + + always = true; + } + else if (strcmp ("close", p) == 0) + { + if (close) + { + cp_parser_error (parser, + "expected modifier % only once"); + cp_parser_skip_to_closing_parenthesis (parser, + /*recovering=*/true, + /*or_comma=*/false, + /*consume_paren=*/true); + return list; } + + close = true; } + else if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type + == CPP_COLON) + { + for (int i = 1; i < pos; ++i) + cp_lexer_consume_token (parser->lexer); + break; + } + else + break; + + if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type == CPP_COMMA) + pos++; + pos++; } if (cp_lexer_next_token_is (parser->lexer, CPP_NAME) diff --git a/gcc/testsuite/c-c++-common/gomp/map-6.c b/gcc/testsuite/c-c++-common/gomp/map-6.c new file mode 100644 index 0000000..f7e22ed --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/map-6.c @@ -0,0 +1,122 @@ +/* { dg-do compile } */ +/* { dg-additional-options "-fdump-tree-original" } */ + +void +foo (void) +{ + /* Test to ensure that the close modifier is parsed and ignored in map clauses. */ + int a, b, b1, b2, b3, b4, b5, b6; + + #pragma omp target map (a) + ; + + #pragma omp target map (to:a) + ; + + #pragma omp target map (a to: b) /* { dg-error "expected '\\)' before 'to'" } */ + ; + + #pragma omp target map (close a) /* { dg-error "'close' undeclared" "" { target c } } */ + /* { dg-error "'close' has not been declared" "" { target c++ } .-1 } */ + /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */ + ; + + #pragma omp target map (always a) /* { dg-error "'always' undeclared" "" { target c } } */ + /* { dg-error "'always' has not been declared" "" { target c++ } .-1 } */ + /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */ + ; + + #pragma omp target map (close to:a) + ; + + #pragma omp target map (close, to:a) + ; + + #pragma omp target map (close delete:a) /* { dg-error "'#pragma omp target' with map-type other than 'to', 'from', 'tofrom' or 'alloc' on 'map' clause" } */ + ; + + #pragma omp target map (close always to:b1) + ; + + #pragma omp target map (close, always to:b2) + ; + + #pragma omp target map (close, always, to:b3) + ; + + #pragma omp target map (always close to:b4) + ; + + #pragma omp target map (always, close to:b5) + ; + + #pragma omp target map (always, close, to:b6) + ; + + #pragma omp target map (always, always, to:a) /* { dg-error "expected modifier 'always' only once" } */ + ; + + #pragma omp target map (always always, to:a) /* { dg-error "expected modifier 'always' only once" } */ + ; + + #pragma omp target map (always, always to:a) /* { dg-error "expected modifier 'always' only once" } */ + ; + + #pragma omp target map (always always to:a) /* { dg-error "expected modifier 'always' only once" } */ + ; + + #pragma omp target map (close, close, to:a) /* { dg-error "expected modifier 'close' only once" } */ + ; + + #pragma omp target map (close close, to:a) /* { dg-error "expected modifier 'close' only once" } */ + ; + + #pragma omp target map (close, close to:a) /* { dg-error "expected modifier 'close' only once" } */ + ; + + #pragma omp target map (close close to:a) /* { dg-error "expected modifier 'close' only once" } */ + ; + + #pragma omp target map (always to : a) map (close to : b) + ; + + int close = 0; + #pragma omp target map (close) + ; + + #pragma omp target map (close a) /* { dg-error "expected '\\)' before 'a'" } */ + ; + + int always = 0; + #pragma omp target map (always) + ; + + #pragma omp target map (always a) /* { dg-error "expected '\\)' before 'a'" } */ + ; + + #pragma omp target map (always, close) + ; + + int to = 0; + #pragma omp target map (always, close, to) + ; + + #pragma omp target map (to, always, close) + { + to = always = close = 1; + } + if (to != 1 || always != 1 || close != 1) + __builtin_abort (); + ; +} + +/* { dg-final { scan-tree-dump-not "map\\(\[^\n\r)]*close\[^\n\r)]*to:" "original" } } */ + +/* { dg-final { scan-tree-dump-times "pragma omp target map\\(always,to:" 6 "original" } } */ + +/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b1" "original" } } */ +/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b2" "original" } } */ +/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b3" "original" } } */ +/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b4" "original" } } */ +/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b5" "original" } } */ +/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b6" "original" } } */ diff --git a/gcc/testsuite/c-c++-common/gomp/map-7.c b/gcc/testsuite/c-c++-common/gomp/map-7.c new file mode 100644 index 0000000..3f1e972 --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/map-7.c @@ -0,0 +1,20 @@ +/* { dg-do compile } */ + +void +foo (void) +{ + /* Test to ensure that the close modifier is parsed and ignored in map clauses. */ + + #define N 1024 + int always[N]; + int close; + + #pragma omp target map(always[:N]) + ; + + #pragma omp target map(close, always[:N]) + ; + + #pragma omp target map(always[:N], close) + ; +}