mirror of
https://gcc.gnu.org/git/gcc.git
synced 2024-11-26 12:23:59 +08:00
OpenMP: 'interop' construct - add C/C++ parser support, improve Fortran parsing
Add middle end support for the 'interop' directive and the 'init', 'use',
and 'destroy' clauses - but fail with a sorry, unimplemented in gimplify.cc.
For Fortran, generate the tree code, update the internal representation,
add some more diagnostic checks and update for newer specification changes
('fr' only takes a single value, but it integer expressions are permitted
again [like with the old syntax] not only constant identifiers).
For C and C++, this patch adds the full parser support for 'interop'.
Still missing is actually handling the directive in the middle end and
in libgomp.
The GOMP_INTEROP_IFR_* internal values have been changed to have space
for vendor specific values that are adjacent to the existing values
but negative, if needed.
gcc/c-family/ChangeLog:
* c-common.h (enum c_omp_region_type): Add C_ORT_INTEROP
and C_ORT_OMP_INTEROP.
(c_omp_interop_t_p): New prototype.
* c-omp.cc (c_omp_interop_t_p): Check whether the type is
omp_interop_t.
(c_omp_directives): Uncomment 'interop'.
* c-pragma.cc (omp_pragmas): Add 'interop'.
* c-pragma.h (enum pragma_kind): Add PRAGMA_OMP_INTEROP.
(enum pragma_omp_clause): Add init, use, and destroy clauses.
gcc/c/ChangeLog:
* c-parser.cc (INCLUDE_STRING): Define.
(c_parser_pragma): Handle 'interop' directive.
(c_parser_omp_clause_name): Handle init, use, and destroy clauses.
(c_parser_omp_all_clauses): Likewise; use C_ORT_OMP_INTEROP, if
'use' is permitted, for c_finish_omp_clauses.
(c_parser_omp_clause_destroy, c_parser_omp_modifier_prefer_type,
c_parser_omp_clause_init, c_parser_omp_clause_use,
OMP_INTEROP_CLAUSE_MASK, c_parser_omp_interop): New.
* c-typeck.cc (c_finish_omp_clauses): Add missing OPT_Wopenmp to
a warning; handle new clauses.
gcc/cp/ChangeLog:
* parser.cc (INCLUDE_STRING): Define.
(cp_parser_omp_clause_name): Handle init, use, and destroy clauses.
(cp_parser_omp_all_clauses): Likewise; use C_ORT_OMP_INTEROP, if
'use' is permitted, for c_finish_omp_clauses.
(cp_parser_omp_modifier_prefer_type, cp_parser_omp_clause_init,
OMP_INTEROP_CLAUSE_MASK, cp_parser_omp_interop): New.
(cp_parser_pragma): Handle 'interop' directive.
* pt.cc (tsubst_omp_clauses): Handle init, use, and destroy clauses.
(tsubst_stmt): Handle OMP_INTEROP.
* semantics.cc (cp_omp_init_prefer_type_update): New.
(finish_omp_clauses): Handle init, use, and destroy clauses
and add clause check for 'depend' on 'interop'.
gcc/fortran/ChangeLog:
* gfortran.h (gfc_omp_namelist): Cleanup interop internal
representation.
* dump-parse-tree.cc (show_omp_namelist): Update for changed
internal representation.
* match.cc (gfc_free_omp_namelist): Likewise.
* openmp.cc (gfc_match_omp_prefer_type, gfc_match_omp_init):
Likewise; also handle some corner cases better and update for
newer 6.0 changes related to 'fr'.
(resolve_omp_clauses): Add type-check for interop variables.
* trans-openmp.cc (gfc_trans_omp_clauses): Handle init, use
and destroy clauses.
(gfc_trans_openmp_interop): New.
(gfc_trans_omp_directive): Call it.
gcc/ChangeLog:
* gimplify.cc (gimplify_expr): Handle OMP_INTEROP by printing
"sorry, uninplemented".
* omp-api.h (omp_get_fr_id_from_name): Change return type to
'char'.
* omp-general.cc (omp_get_fr_id_from_name): Likewise; return
GOMP_INTEROP_IFR_UNKNOWN not 0 if not found.
(omp_get_name_from_fr_id): Return "<unknown>" not NULL
if not found (used for dumps).
* tree-core.h (enum omp_clause_code): Add OMP_CLAUSE_DESTROY,
OMP_CLAUSE_USE, and OMP_CLAUSE_INIT.
* tree-pretty-print.cc (dump_omp_init_prefer_type): New.
(dump_omp_clause): Handle init, use and destroy clauses.
(dump_generic_node): Handle interop directive.
* tree.cc (omp_clause_num_ops, omp_clause_code_name): Add new
init/use/destroy clauses.
* tree.def (OACC_LOOP): Fix comment.
(OMP_INTEROP): Add.
* tree.h (OMP_INTEROP_CLAUSES, OMP_CLAUSE_INIT_TARGET,
OMP_CLAUSE_INIT_TARGETSYNC, OMP_CLAUSE_INIT_PREFER_TYPE): New.
include/ChangeLog:
* gomp-constants.h (GOMP_INTEROP_IFR_NONE): Rename ...
(GOMP_INTEROP_IFR_UNKNOWN): ... to this. And change value.
(GOMP_INTEROP_IFR_SEPARATOR): Likewise.
gcc/testsuite/ChangeLog:
* gfortran.dg/gomp/interop-1.f90: Update for parser changes,
spec changes and add new tests.
* gfortran.dg/gomp/interop-2.f90: Likewise.
* gfortran.dg/gomp/interop-3.f90: Likewise.
* c-c++-common/gomp/interop-1.c: New test.
* c-c++-common/gomp/interop-2.c: New test.
* c-c++-common/gomp/interop-3.c: New test.
* c-c++-common/gomp/interop-4.c: New test.
* g++.dg/gomp/interop-5.C: New test.
* gfortran.dg/gomp/interop-4.f90: New test.
This commit is contained in:
Tobias Burnus
parent
8d7f2d53c8
commit
8f0c8e577a
@ -1296,9 +1296,11 @@ enum c_omp_region_type
|
||||
C_ORT_DECLARE_SIMD = 1 << 2,
|
||||
C_ORT_TARGET = 1 << 3,
|
||||
C_ORT_EXIT_DATA = 1 << 4,
|
||||
C_ORT_INTEROP = 1 << 5,
|
||||
C_ORT_OMP_DECLARE_SIMD = C_ORT_OMP | C_ORT_DECLARE_SIMD,
|
||||
C_ORT_OMP_TARGET = C_ORT_OMP | C_ORT_TARGET,
|
||||
C_ORT_OMP_EXIT_DATA = C_ORT_OMP | C_ORT_EXIT_DATA,
|
||||
C_ORT_OMP_INTEROP = C_ORT_OMP | C_ORT_INTEROP,
|
||||
C_ORT_ACC_TARGET = C_ORT_ACC | C_ORT_TARGET
|
||||
};
|
||||
|
||||
@ -1311,6 +1313,7 @@ extern void c_finish_omp_barrier (location_t);
|
||||
extern tree c_finish_omp_atomic (location_t, enum tree_code, enum tree_code,
|
||||
tree, tree, tree, tree, tree, tree, bool,
|
||||
enum omp_memory_order, bool, bool = false);
|
||||
extern bool c_omp_interop_t_p (tree);
|
||||
extern bool c_omp_depend_t_p (tree);
|
||||
extern void c_finish_omp_depobj (location_t, tree, enum omp_clause_depend_kind,
|
||||
tree);
|
||||
|
||||
@ -664,6 +664,24 @@ c_finish_omp_atomic (location_t loc, enum tree_code code,
|
||||
}
|
||||
|
||||
|
||||
/* Return true if TYPE is the implementation's omp_interop_t. */
|
||||
|
||||
bool
|
||||
c_omp_interop_t_p (tree type)
|
||||
{
|
||||
type = TYPE_MAIN_VARIANT (type);
|
||||
return (TREE_CODE (type) == ENUMERAL_TYPE
|
||||
&& TYPE_NAME (type)
|
||||
&& ((TREE_CODE (TYPE_NAME (type)) == TYPE_DECL
|
||||
? DECL_NAME (TYPE_NAME (type)) : TYPE_NAME (type))
|
||||
== get_identifier ("omp_interop_t"))
|
||||
&& TYPE_FILE_SCOPE_P (type)
|
||||
&& COMPLETE_TYPE_P (type)
|
||||
&& TREE_CODE (TYPE_SIZE (type)) == INTEGER_CST
|
||||
&& !compare_tree_int (TYPE_SIZE (type),
|
||||
tree_to_uhwi (TYPE_SIZE (ptr_type_node))));
|
||||
}
|
||||
|
||||
/* Return true if TYPE is the implementation's omp_depend_t. */
|
||||
|
||||
bool
|
||||
@ -4321,8 +4339,8 @@ const struct c_omp_directive c_omp_directives[] = {
|
||||
C_OMP_DIR_CONSTRUCT, true },
|
||||
/* { "groupprivate", nullptr, nullptr, PRAGMA_OMP_GROUPPRIVATE,
|
||||
C_OMP_DIR_DECLARATIVE, false }, */
|
||||
/* { "interop", nullptr, nullptr, PRAGMA_OMP_INTEROP,
|
||||
C_OMP_DIR_STANDALONE, false }, */
|
||||
{ "interop", nullptr, nullptr, PRAGMA_OMP_INTEROP,
|
||||
C_OMP_DIR_STANDALONE, false },
|
||||
{ "loop", nullptr, nullptr, PRAGMA_OMP_LOOP,
|
||||
C_OMP_DIR_CONSTRUCT, true },
|
||||
{ "masked", nullptr, nullptr, PRAGMA_OMP_MASKED,
|
||||
|
||||
@ -1529,6 +1529,7 @@ static const struct omp_pragma_def omp_pragmas[] = {
|
||||
{ "error", PRAGMA_OMP_ERROR },
|
||||
{ "end", PRAGMA_OMP_END },
|
||||
{ "flush", PRAGMA_OMP_FLUSH },
|
||||
{ "interop", PRAGMA_OMP_INTEROP },
|
||||
{ "nothing", PRAGMA_OMP_NOTHING },
|
||||
{ "requires", PRAGMA_OMP_REQUIRES },
|
||||
{ "scope", PRAGMA_OMP_SCOPE },
|
||||
|
||||
@ -61,6 +61,7 @@ enum pragma_kind {
|
||||
PRAGMA_OMP_END,
|
||||
PRAGMA_OMP_FLUSH,
|
||||
PRAGMA_OMP_FOR,
|
||||
PRAGMA_OMP_INTEROP,
|
||||
PRAGMA_OMP_LOOP,
|
||||
PRAGMA_OMP_NOTHING,
|
||||
PRAGMA_OMP_MASKED,
|
||||
@ -111,6 +112,7 @@ enum pragma_omp_clause {
|
||||
PRAGMA_OMP_CLAUSE_DEFAULT,
|
||||
PRAGMA_OMP_CLAUSE_DEFAULTMAP,
|
||||
PRAGMA_OMP_CLAUSE_DEPEND,
|
||||
PRAGMA_OMP_CLAUSE_DESTROY,
|
||||
PRAGMA_OMP_CLAUSE_DETACH,
|
||||
PRAGMA_OMP_CLAUSE_DEVICE,
|
||||
PRAGMA_OMP_CLAUSE_DEVICE_TYPE,
|
||||
@ -130,6 +132,7 @@ enum pragma_omp_clause {
|
||||
PRAGMA_OMP_CLAUSE_IN_REDUCTION,
|
||||
PRAGMA_OMP_CLAUSE_INBRANCH,
|
||||
PRAGMA_OMP_CLAUSE_INDIRECT,
|
||||
PRAGMA_OMP_CLAUSE_INIT,
|
||||
PRAGMA_OMP_CLAUSE_IS_DEVICE_PTR,
|
||||
PRAGMA_OMP_CLAUSE_LASTPRIVATE,
|
||||
PRAGMA_OMP_CLAUSE_LINEAR,
|
||||
@ -166,6 +169,7 @@ enum pragma_omp_clause {
|
||||
PRAGMA_OMP_CLAUSE_TO,
|
||||
PRAGMA_OMP_CLAUSE_UNIFORM,
|
||||
PRAGMA_OMP_CLAUSE_UNTIED,
|
||||
PRAGMA_OMP_CLAUSE_USE,
|
||||
PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR,
|
||||
PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR,
|
||||
|
||||
|
||||
@ -37,6 +37,7 @@ along with GCC; see the file COPYING3. If not see
|
||||
|
||||
#include "config.h"
|
||||
#define INCLUDE_MEMORY
|
||||
#define INCLUDE_STRING
|
||||
#include "system.h"
|
||||
#include "coretypes.h"
|
||||
#include "target.h"
|
||||
@ -1753,6 +1754,7 @@ static void c_parser_omp_allocate (c_parser *);
|
||||
static void c_parser_omp_assumes (c_parser *);
|
||||
static bool c_parser_omp_ordered (c_parser *, enum pragma_context, bool *);
|
||||
static tree c_parser_omp_dispatch (location_t, c_parser *);
|
||||
static void c_parser_omp_interop (c_parser *);
|
||||
static void c_parser_oacc_routine (c_parser *, enum pragma_context);
|
||||
|
||||
/* These Objective-C parser functions are only ever called when
|
||||
@ -15384,6 +15386,15 @@ c_parser_pragma (c_parser *parser, enum pragma_context context, bool *if_p,
|
||||
c_parser_omp_flush (parser);
|
||||
return false;
|
||||
|
||||
case PRAGMA_OMP_INTEROP:
|
||||
if (context != pragma_compound)
|
||||
{
|
||||
construct = "omp interop";
|
||||
goto in_compound;
|
||||
}
|
||||
c_parser_omp_interop (parser);
|
||||
return false;
|
||||
|
||||
case PRAGMA_OMP_TASKWAIT:
|
||||
if (context != pragma_compound)
|
||||
{
|
||||
@ -15711,6 +15722,8 @@ c_parser_omp_clause_name (c_parser *parser)
|
||||
result = PRAGMA_OACC_CLAUSE_DELETE;
|
||||
else if (!strcmp ("depend", p))
|
||||
result = PRAGMA_OMP_CLAUSE_DEPEND;
|
||||
else if (!strcmp ("destroy", p))
|
||||
result = PRAGMA_OMP_CLAUSE_DESTROY;
|
||||
else if (!strcmp ("detach", p))
|
||||
result = PRAGMA_OACC_CLAUSE_DETACH;
|
||||
else if (!strcmp ("device", p))
|
||||
@ -15769,6 +15782,8 @@ c_parser_omp_clause_name (c_parser *parser)
|
||||
result = PRAGMA_OACC_CLAUSE_INDEPENDENT;
|
||||
else if (!strcmp ("indirect", p))
|
||||
result = PRAGMA_OMP_CLAUSE_INDIRECT;
|
||||
else if (!strcmp ("init", p))
|
||||
result = PRAGMA_OMP_CLAUSE_INIT;
|
||||
else if (!strcmp ("is_device_ptr", p))
|
||||
result = PRAGMA_OMP_CLAUSE_IS_DEVICE_PTR;
|
||||
break;
|
||||
@ -15889,6 +15904,8 @@ c_parser_omp_clause_name (c_parser *parser)
|
||||
result = PRAGMA_OMP_CLAUSE_UNIFORM;
|
||||
else if (!strcmp ("untied", p))
|
||||
result = PRAGMA_OMP_CLAUSE_UNTIED;
|
||||
else if (!strcmp ("use", p))
|
||||
result = PRAGMA_OMP_CLAUSE_USE;
|
||||
else if (!strcmp ("use_device", p))
|
||||
result = PRAGMA_OACC_CLAUSE_USE_DEVICE;
|
||||
else if (!strcmp ("use_device_addr", p))
|
||||
@ -20172,6 +20189,386 @@ c_parser_omp_clause_detach (c_parser *parser, tree list)
|
||||
return u;
|
||||
}
|
||||
|
||||
/* OpenMP 5.0:
|
||||
destroy ( variable-list ) */
|
||||
|
||||
static tree
|
||||
c_parser_omp_clause_destroy (c_parser *parser, tree list)
|
||||
{
|
||||
return c_parser_omp_var_list_parens (parser, OMP_CLAUSE_DESTROY, list);
|
||||
}
|
||||
|
||||
/* OpenMP 5.1:
|
||||
prefer_type ( const-int-expr-or-string-literal-list )
|
||||
|
||||
OpenMP 6.0:
|
||||
prefer_type ( { preference-selector-list }, { ... } )
|
||||
|
||||
with preference-selector being:
|
||||
fr ( identifier-or-string-literal-list )
|
||||
attr ( string-list )
|
||||
|
||||
Data format:
|
||||
For the foreign runtime identifiers, string values are converted to
|
||||
their integer value; unknown string or integer values are set to
|
||||
GOMP_INTEROP_IFR_KNOWN.
|
||||
|
||||
Each item (a) GOMP_INTEROP_IFR_SEPARATOR
|
||||
(b) for any 'fr', its integer value.
|
||||
Note: Spec only permits 1 'fr' entry (6.0; changed after TR13)
|
||||
(c) GOMP_INTEROP_IFR_SEPARATOR
|
||||
(d) list of \0-terminated non-empty strings for 'attr'
|
||||
(e) '\0'
|
||||
Tailing '\0'. */
|
||||
|
||||
static tree
|
||||
c_parser_omp_modifier_prefer_type (c_parser *parser)
|
||||
{
|
||||
if (!c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>"))
|
||||
return error_mark_node;
|
||||
|
||||
std::string str;
|
||||
|
||||
/* Old Format: const-int-expr-or-string-literal-list */
|
||||
if (!c_parser_next_token_is (parser, CPP_OPEN_BRACE))
|
||||
while (true)
|
||||
{
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
if (c_parser_next_token_is (parser, CPP_STRING))
|
||||
{
|
||||
c_expr cval = c_parser_string_literal (parser, false, false);
|
||||
if (cval.value == error_mark_node)
|
||||
return error_mark_node;
|
||||
if ((size_t) TREE_STRING_LENGTH (cval.value)
|
||||
!= strlen (TREE_STRING_POINTER (cval.value)) + 1)
|
||||
{
|
||||
error_at (cval.get_location (), "string literal must "
|
||||
"not contain %<\\0%>");
|
||||
parser->error = true;
|
||||
return error_mark_node;
|
||||
}
|
||||
|
||||
char c = omp_get_fr_id_from_name (TREE_STRING_POINTER (cval.value));
|
||||
if (c == GOMP_INTEROP_IFR_UNKNOWN)
|
||||
warning_at (cval.get_location (), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qs",
|
||||
TREE_STRING_POINTER (cval.value));
|
||||
str += c;
|
||||
}
|
||||
else
|
||||
{
|
||||
c_expr cval = c_parser_expr_no_commas (parser, NULL);
|
||||
tree value = c_fully_fold (cval.value, false, NULL);
|
||||
if (INTEGRAL_TYPE_P (TREE_TYPE (value))
|
||||
&& TREE_CODE (value) != INTEGER_CST)
|
||||
value = convert_lvalue_to_rvalue (cval.get_start (), cval,
|
||||
false, true).value;
|
||||
if (TREE_CODE (value) != INTEGER_CST
|
||||
|| !tree_fits_shwi_p (value))
|
||||
{
|
||||
c_parser_error (parser, "expected string literal or constant "
|
||||
"integer expression");
|
||||
return error_mark_node;
|
||||
}
|
||||
HOST_WIDE_INT n = tree_to_shwi (value);
|
||||
if (n < 1 || n > GOMP_INTEROP_IFR_LAST)
|
||||
{
|
||||
warning_at (cval.get_location (), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qwd", n);
|
||||
n = GOMP_INTEROP_IFR_UNKNOWN;
|
||||
}
|
||||
str += (char) n;
|
||||
}
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
str += '\0';
|
||||
if (c_parser_next_token_is (parser, CPP_COMMA))
|
||||
{
|
||||
c_parser_consume_token (parser);
|
||||
continue;
|
||||
}
|
||||
if (!c_parser_require (parser, CPP_CLOSE_PAREN,
|
||||
"expected %<,%> or %<)%>"))
|
||||
return error_mark_node;
|
||||
str += '\0';
|
||||
tree res = build_string (str.length (), str.data ());
|
||||
TREE_TYPE (res) = build_array_type_nelts (unsigned_char_type_node,
|
||||
str.length ());
|
||||
return res;
|
||||
}
|
||||
|
||||
/* New format. */
|
||||
std::string str2;
|
||||
while (true)
|
||||
{
|
||||
if (!c_parser_require (parser, CPP_OPEN_BRACE, "expected %<{%>"))
|
||||
return error_mark_node;
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
str2.clear ();
|
||||
bool has_fr = false;
|
||||
while (true)
|
||||
{
|
||||
c_token *tok = c_parser_peek_token (parser);
|
||||
if (tok->type != CPP_NAME
|
||||
|| (strcmp("fr", IDENTIFIER_POINTER (tok->value)) != 0
|
||||
&& strcmp("attr", IDENTIFIER_POINTER (tok->value)) != 0))
|
||||
{
|
||||
c_parser_error (parser, "expected %<fr%> or %<attr%> preference "
|
||||
"selector");
|
||||
return error_mark_node;
|
||||
}
|
||||
c_parser_consume_token (parser);
|
||||
bool is_fr = IDENTIFIER_POINTER (tok->value)[0] == 'f';
|
||||
if (is_fr && has_fr)
|
||||
{
|
||||
c_parser_error (parser, "duplicated %<fr%> preference selector");
|
||||
return error_mark_node;
|
||||
}
|
||||
if (!c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>"))
|
||||
return error_mark_node;
|
||||
while (true)
|
||||
{
|
||||
if (c_parser_next_token_is (parser, CPP_STRING))
|
||||
{
|
||||
c_expr cval = c_parser_string_literal (parser, false, false);
|
||||
tree value = cval.value;
|
||||
if (value == error_mark_node)
|
||||
return error_mark_node;
|
||||
if ((size_t) TREE_STRING_LENGTH (value)
|
||||
!= strlen (TREE_STRING_POINTER (value)) + 1)
|
||||
{
|
||||
error_at (cval.get_location (), "string literal must "
|
||||
"not contain %<\\0%>");
|
||||
parser->error = true;
|
||||
return error_mark_node;
|
||||
}
|
||||
if (!is_fr)
|
||||
{
|
||||
if (!startswith (TREE_STRING_POINTER (value), "ompx_"))
|
||||
{
|
||||
error_at (cval.get_location (),
|
||||
"%<attr%> string literal must start with "
|
||||
"%<ompx_%>");
|
||||
parser->error = true;
|
||||
return error_mark_node;
|
||||
}
|
||||
if (strchr (TREE_STRING_POINTER (value), ','))
|
||||
{
|
||||
error_at (cval.get_location (),
|
||||
"%<attr%> string literal must not contain "
|
||||
"a comma");
|
||||
parser->error = true;
|
||||
return error_mark_node;
|
||||
}
|
||||
str2 += TREE_STRING_POINTER (value);
|
||||
str2 += '\0';
|
||||
}
|
||||
else
|
||||
{
|
||||
if (*TREE_STRING_POINTER (value) == '\0')
|
||||
{
|
||||
c_parser_error (parser, "non-empty string literal expected");
|
||||
return error_mark_node;
|
||||
}
|
||||
char c = omp_get_fr_id_from_name (TREE_STRING_POINTER (value));
|
||||
if (c == GOMP_INTEROP_IFR_UNKNOWN)
|
||||
warning_at (cval.get_location (), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qs",
|
||||
TREE_STRING_POINTER (value));
|
||||
str += c;
|
||||
has_fr = true;
|
||||
}
|
||||
}
|
||||
else if (!is_fr)
|
||||
{
|
||||
c_parser_error (parser, "expected string literal");
|
||||
return error_mark_node;
|
||||
}
|
||||
else
|
||||
{
|
||||
c_expr cval = c_parser_expr_no_commas (parser, NULL);
|
||||
tree value = c_fully_fold (cval.value, false, NULL);
|
||||
if (INTEGRAL_TYPE_P (TREE_TYPE (value))
|
||||
&& TREE_CODE (value) != INTEGER_CST)
|
||||
value = convert_lvalue_to_rvalue (cval.get_start (), cval,
|
||||
false, true).value;
|
||||
|
||||
if (TREE_CODE (value) != INTEGER_CST
|
||||
|| !tree_fits_shwi_p (value))
|
||||
{
|
||||
c_parser_error (parser, "expected string literal or "
|
||||
"constant integer expression");
|
||||
return error_mark_node;
|
||||
}
|
||||
HOST_WIDE_INT n = tree_to_shwi (value);
|
||||
if (n < 1 || n > GOMP_INTEROP_IFR_LAST)
|
||||
{
|
||||
warning_at (cval.get_location (), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qwd", n);
|
||||
n = GOMP_INTEROP_IFR_UNKNOWN;
|
||||
}
|
||||
str += (char) n;
|
||||
has_fr = true;
|
||||
}
|
||||
if (!is_fr
|
||||
&& c_parser_next_token_is (parser, CPP_COMMA))
|
||||
{
|
||||
c_parser_consume_token (parser);
|
||||
continue;
|
||||
}
|
||||
if (!c_parser_require (parser, CPP_CLOSE_PAREN,
|
||||
is_fr ? G_("expected %<)%>")
|
||||
: G_("expected %<)%> or %<,%>")))
|
||||
return error_mark_node;
|
||||
break;
|
||||
}
|
||||
if (c_parser_next_token_is (parser, CPP_COMMA))
|
||||
{
|
||||
c_parser_consume_token (parser);
|
||||
continue;
|
||||
}
|
||||
if (c_parser_next_token_is (parser, CPP_CLOSE_BRACE))
|
||||
break;
|
||||
c_parser_error (parser, "expected %<,%> or %<}%>");
|
||||
return error_mark_node;
|
||||
}
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
str += str2;
|
||||
str += '\0';
|
||||
c_parser_consume_token (parser);
|
||||
if (c_parser_next_token_is (parser, CPP_CLOSE_PAREN))
|
||||
break;
|
||||
if (!c_parser_require (parser, CPP_COMMA, "expected %<)%> or %<,%>"))
|
||||
return error_mark_node;
|
||||
}
|
||||
c_parser_consume_token (parser);
|
||||
str += '\0';
|
||||
tree res = build_string (str.length (), str.data ());
|
||||
TREE_TYPE (res) = build_array_type_nelts (unsigned_char_type_node,
|
||||
str.length ());
|
||||
return res;
|
||||
}
|
||||
|
||||
/* OpenMP 5.1:
|
||||
init ( [init-modifier-list : ] variable-list )
|
||||
|
||||
Modifiers:
|
||||
target
|
||||
targetsync
|
||||
prefer_type (preference-specification) */
|
||||
|
||||
static tree
|
||||
c_parser_omp_clause_init (c_parser *parser, tree list)
|
||||
{
|
||||
location_t loc = c_parser_peek_token (parser)->location;
|
||||
|
||||
matching_parens parens;
|
||||
if (!parens.require_open (parser))
|
||||
return list;
|
||||
|
||||
unsigned pos = 0, raw_pos = 1;
|
||||
while (c_parser_peek_nth_token_raw (parser, raw_pos)->type == CPP_NAME)
|
||||
{
|
||||
pos++; raw_pos++;
|
||||
if (c_parser_peek_nth_token_raw (parser, raw_pos)->type == CPP_OPEN_PAREN)
|
||||
{
|
||||
raw_pos++;
|
||||
c_parser_check_balanced_raw_token_sequence (parser, &raw_pos);
|
||||
if (c_parser_peek_nth_token_raw (parser, raw_pos)->type != CPP_CLOSE_PAREN)
|
||||
{
|
||||
pos = 0;
|
||||
break;
|
||||
}
|
||||
raw_pos++;
|
||||
}
|
||||
if (c_parser_peek_nth_token_raw (parser, raw_pos)->type == CPP_COLON)
|
||||
break;
|
||||
if (c_parser_peek_nth_token_raw (parser, raw_pos)->type != CPP_COMMA)
|
||||
{
|
||||
pos = 0;
|
||||
break;
|
||||
}
|
||||
pos++;
|
||||
raw_pos++;
|
||||
}
|
||||
|
||||
bool target = false;
|
||||
bool targetsync = false;
|
||||
tree prefer_type_tree = NULL_TREE;
|
||||
|
||||
for (unsigned pos2 = 0; pos2 < pos; ++pos2)
|
||||
{
|
||||
c_token *tok = c_parser_peek_token (parser);
|
||||
if (tok->type == CPP_COMMA)
|
||||
{
|
||||
c_parser_consume_token (parser);
|
||||
continue;
|
||||
}
|
||||
|
||||
const char *p = IDENTIFIER_POINTER (tok->value);
|
||||
if (strcmp ("targetsync", p) == 0)
|
||||
{
|
||||
if (targetsync)
|
||||
error_at (tok->location, "duplicate %<targetsync%> modifier");
|
||||
targetsync = true;
|
||||
c_parser_consume_token (parser);
|
||||
}
|
||||
else if (strcmp ("target", p) == 0)
|
||||
{
|
||||
if (target)
|
||||
error_at (tok->location, "duplicate %<target%> modifier");
|
||||
target = true;
|
||||
c_parser_consume_token (parser);
|
||||
}
|
||||
else if (strcmp ("prefer_type", p) == 0)
|
||||
{
|
||||
if (prefer_type_tree != NULL_TREE)
|
||||
error_at (tok->location, "duplicate %<prefer_type%> modifier");
|
||||
c_parser_consume_token (parser);
|
||||
prefer_type_tree = c_parser_omp_modifier_prefer_type (parser);
|
||||
if (prefer_type_tree == error_mark_node)
|
||||
return list;
|
||||
}
|
||||
else
|
||||
{
|
||||
c_parser_error (parser, "%<init%> clause with modifier other than "
|
||||
"%<prefer_type%>, %<target%> or "
|
||||
"%<targetsync%>");
|
||||
parens.skip_until_found_close (parser);
|
||||
return list;
|
||||
}
|
||||
}
|
||||
if (pos)
|
||||
{
|
||||
c_token *tok = c_parser_peek_token (parser);
|
||||
gcc_checking_assert (tok->type == CPP_COLON);
|
||||
c_parser_consume_token (parser);
|
||||
}
|
||||
|
||||
tree nl = c_parser_omp_variable_list (parser, loc, OMP_CLAUSE_INIT, list,
|
||||
false);
|
||||
parens.skip_until_found_close (parser);
|
||||
|
||||
for (tree c = nl; c != list; c = OMP_CLAUSE_CHAIN (c))
|
||||
{
|
||||
if (target)
|
||||
OMP_CLAUSE_INIT_TARGET (c) = 1;
|
||||
if (targetsync)
|
||||
OMP_CLAUSE_INIT_TARGETSYNC (c) = 1;
|
||||
if (prefer_type_tree)
|
||||
OMP_CLAUSE_INIT_PREFER_TYPE (c) = prefer_type_tree;
|
||||
}
|
||||
return nl;
|
||||
}
|
||||
|
||||
/* OpenMP 5.0:
|
||||
use ( variable-list ) */
|
||||
|
||||
static tree
|
||||
c_parser_omp_clause_use (c_parser *parser, tree list)
|
||||
{
|
||||
return c_parser_omp_var_list_parens (parser, OMP_CLAUSE_USE, list);
|
||||
}
|
||||
|
||||
/* Parse all OpenACC clauses. The set clauses allowed by the directive
|
||||
is a bitmask in MASK. Return the list of clauses found. */
|
||||
|
||||
@ -20667,6 +21064,18 @@ c_parser_omp_all_clauses (c_parser *parser, omp_clause_mask mask,
|
||||
clauses = c_parser_omp_clause_doacross (parser, clauses);
|
||||
c_name = "doacross";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_DESTROY:
|
||||
clauses = c_parser_omp_clause_destroy (parser, clauses);
|
||||
c_name = "destroy";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_INIT:
|
||||
clauses = c_parser_omp_clause_init (parser, clauses);
|
||||
c_name = "init";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_USE:
|
||||
clauses = c_parser_omp_clause_use (parser, clauses);
|
||||
c_name = "use";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_MAP:
|
||||
clauses = c_parser_omp_clause_map (parser, clauses);
|
||||
c_name = "map";
|
||||
@ -20773,6 +21182,8 @@ c_parser_omp_all_clauses (c_parser *parser, omp_clause_mask mask,
|
||||
{
|
||||
if ((mask & (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_UNIFORM)) != 0)
|
||||
return c_finish_omp_clauses (clauses, C_ORT_OMP_DECLARE_SIMD);
|
||||
if ((mask & (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_USE)) != 0)
|
||||
return c_finish_omp_clauses (clauses, C_ORT_OMP_INTEROP);
|
||||
return c_finish_omp_clauses (clauses, C_ORT_OMP);
|
||||
}
|
||||
|
||||
@ -24109,6 +24520,32 @@ c_parser_omp_masked (location_t loc, c_parser *parser,
|
||||
clauses);
|
||||
}
|
||||
|
||||
/* OpenMP 5.1:
|
||||
# pragma omp interop clauses[opt] new-line */
|
||||
|
||||
#define OMP_INTEROP_CLAUSE_MASK \
|
||||
( (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DEPEND) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DESTROY) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DEVICE) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_INIT) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_NOWAIT) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_USE))
|
||||
|
||||
static void
|
||||
c_parser_omp_interop (c_parser *parser)
|
||||
{
|
||||
location_t loc = c_parser_peek_token (parser)->location;
|
||||
c_parser_consume_pragma (parser);
|
||||
tree clauses = c_parser_omp_all_clauses (parser,
|
||||
OMP_INTEROP_CLAUSE_MASK,
|
||||
"#pragma omp interop");
|
||||
tree stmt = make_node (OMP_INTEROP);
|
||||
TREE_TYPE (stmt) = void_type_node;
|
||||
OMP_INTEROP_CLAUSES (stmt) = clauses;
|
||||
SET_EXPR_LOCATION (stmt, loc);
|
||||
add_stmt (stmt);
|
||||
}
|
||||
|
||||
/* OpenMP 2.5:
|
||||
# pragma omp ordered new-line
|
||||
structured-block
|
||||
|
||||
@ -15670,6 +15670,10 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
tree *full_seen = NULL;
|
||||
bool partial_seen = false;
|
||||
bool openacc = (ort & C_ORT_ACC) != 0;
|
||||
bool init_seen = false;
|
||||
bool init_use_destroy_seen = false;
|
||||
tree init_no_targetsync_clause = NULL_TREE;
|
||||
tree depend_clause = NULL_TREE;
|
||||
|
||||
bitmap_obstack_initialize (NULL);
|
||||
bitmap_initialize (&generic_head, &bitmap_default_obstack);
|
||||
@ -16292,7 +16296,7 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
}
|
||||
else if (bitmap_bit_p (&aligned_head, DECL_UID (t)))
|
||||
{
|
||||
warning_at (OMP_CLAUSE_LOCATION (c), 0,
|
||||
warning_at (OMP_CLAUSE_LOCATION (c), OPT_Wopenmp,
|
||||
"%qE appears more than once in %<allocate%> clauses",
|
||||
t);
|
||||
remove = true;
|
||||
@ -16339,6 +16343,8 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
}
|
||||
gcc_unreachable ();
|
||||
case OMP_CLAUSE_DEPEND:
|
||||
depend_clause = c;
|
||||
/* FALLTHRU */
|
||||
case OMP_CLAUSE_AFFINITY:
|
||||
t = OMP_CLAUSE_DECL (c);
|
||||
if (TREE_CODE (t) == TREE_LIST
|
||||
@ -17071,6 +17077,32 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
}
|
||||
break;
|
||||
|
||||
case OMP_CLAUSE_INIT:
|
||||
init_seen = true;
|
||||
if (!OMP_CLAUSE_INIT_TARGETSYNC (c))
|
||||
init_no_targetsync_clause = c;
|
||||
/* FALLTHRU */
|
||||
case OMP_CLAUSE_DESTROY:
|
||||
case OMP_CLAUSE_USE:
|
||||
init_use_destroy_seen = true;
|
||||
t = OMP_CLAUSE_DECL (c);
|
||||
if (bitmap_bit_p (&generic_head, DECL_UID (t)))
|
||||
{
|
||||
error_at (OMP_CLAUSE_LOCATION (c),
|
||||
"%qD appears more than once in action clauses", t);
|
||||
remove = true;
|
||||
}
|
||||
else if (/* ort == C_ORT_OMP_INTEROP [uncomment for depobj init] */
|
||||
!c_omp_interop_t_p (TREE_TYPE (OMP_CLAUSE_DECL (c))))
|
||||
error_at (OMP_CLAUSE_LOCATION (c),
|
||||
"%qD must be of %<omp_interop_t%>", OMP_CLAUSE_DECL (c));
|
||||
else if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_USE
|
||||
&& TREE_READONLY (OMP_CLAUSE_DECL (c)))
|
||||
error_at (OMP_CLAUSE_LOCATION (c),
|
||||
"%qD shall not be const", OMP_CLAUSE_DECL (c));
|
||||
bitmap_set_bit (&generic_head, DECL_UID (t));
|
||||
pc = &OMP_CLAUSE_CHAIN (c);
|
||||
break;
|
||||
default:
|
||||
gcc_unreachable ();
|
||||
}
|
||||
@ -17342,6 +17374,19 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
}
|
||||
}
|
||||
|
||||
if (ort == C_ORT_OMP_INTEROP
|
||||
&& depend_clause
|
||||
&& (!init_use_destroy_seen
|
||||
|| (init_seen && init_no_targetsync_clause)))
|
||||
{
|
||||
error_at (OMP_CLAUSE_LOCATION (depend_clause),
|
||||
"%<depend%> clause requires action clauses with "
|
||||
"%<targetsync%> interop-type");
|
||||
if (init_no_targetsync_clause)
|
||||
inform (OMP_CLAUSE_LOCATION (init_no_targetsync_clause),
|
||||
"%<init%> clause lacks the %<targetsync%> modifier");
|
||||
}
|
||||
|
||||
bitmap_obstack_release (NULL);
|
||||
return clauses;
|
||||
}
|
||||
|
||||
485
gcc/cp/parser.cc
485
gcc/cp/parser.cc
@ -21,6 +21,7 @@ along with GCC; see the file COPYING3. If not see
|
||||
#include "config.h"
|
||||
#include "omp-selectors.h"
|
||||
#define INCLUDE_MEMORY
|
||||
#define INCLUDE_STRING
|
||||
#include "system.h"
|
||||
#include "coretypes.h"
|
||||
#include "cp-tree.h"
|
||||
@ -38272,6 +38273,8 @@ cp_parser_omp_clause_name (cp_parser *parser)
|
||||
result = PRAGMA_OMP_CLAUSE_DEFAULTMAP;
|
||||
else if (!strcmp ("depend", p))
|
||||
result = PRAGMA_OMP_CLAUSE_DEPEND;
|
||||
else if (!strcmp ("destroy", p))
|
||||
result = PRAGMA_OMP_CLAUSE_DESTROY;
|
||||
else if (!strcmp ("detach", p))
|
||||
result = PRAGMA_OACC_CLAUSE_DETACH;
|
||||
else if (!strcmp ("device", p))
|
||||
@ -38330,6 +38333,8 @@ cp_parser_omp_clause_name (cp_parser *parser)
|
||||
result = PRAGMA_OACC_CLAUSE_INDEPENDENT;
|
||||
else if (!strcmp ("indirect", p))
|
||||
result = PRAGMA_OMP_CLAUSE_INDIRECT;
|
||||
else if (!strcmp ("init", p))
|
||||
result = PRAGMA_OMP_CLAUSE_INIT;
|
||||
else if (!strcmp ("is_device_ptr", p))
|
||||
result = PRAGMA_OMP_CLAUSE_IS_DEVICE_PTR;
|
||||
break;
|
||||
@ -38446,6 +38451,8 @@ cp_parser_omp_clause_name (cp_parser *parser)
|
||||
result = PRAGMA_OMP_CLAUSE_UNIFORM;
|
||||
else if (!strcmp ("untied", p))
|
||||
result = PRAGMA_OMP_CLAUSE_UNTIED;
|
||||
else if (!strcmp ("use", p))
|
||||
result = PRAGMA_OMP_CLAUSE_USE;
|
||||
else if (!strcmp ("use_device", p))
|
||||
result = PRAGMA_OACC_CLAUSE_USE_DEVICE;
|
||||
else if (!strcmp ("use_device_addr", p))
|
||||
@ -42271,6 +42278,418 @@ cp_parser_omp_clause_device_type (cp_parser *parser, tree list,
|
||||
return list;
|
||||
}
|
||||
|
||||
/* OpenMP 5.1:
|
||||
prefer_type ( const-int-expr-or-string-literal-list )
|
||||
|
||||
OpenMP 6.0:
|
||||
prefer_type ( { preference-selector-list }, { ... } )
|
||||
|
||||
with preference-selector being:
|
||||
fr ( identifier-or-string-literal-list )
|
||||
attr ( string-list )
|
||||
|
||||
Data format:
|
||||
For the foreign runtime identifiers, string values are converted to
|
||||
their integer value; unknown string or integer values are set to
|
||||
GOMP_INTEROP_IFR_KNOWN.
|
||||
|
||||
Each item (a) GOMP_INTEROP_IFR_SEPARATOR
|
||||
(b) for any 'fr', its integer value.
|
||||
Note: Spec only permits 1 'fr' entry (6.0; changed after TR13)
|
||||
(c) GOMP_INTEROP_IFR_SEPARATOR
|
||||
(d) list of \0-terminated non-empty strings for 'attr'
|
||||
(e) '\0'
|
||||
Tailing '\0'.
|
||||
|
||||
When processing a template:
|
||||
attr and fr strings are processed normally.
|
||||
for integer expressions, set fr to UNKNOWN and keep a separate list
|
||||
of those expressions - and store it as
|
||||
tree_cons (bytestring, fr_tree list). */
|
||||
|
||||
static tree
|
||||
cp_parser_omp_modifier_prefer_type (cp_parser *parser)
|
||||
{
|
||||
if (!cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN))
|
||||
return error_mark_node;
|
||||
|
||||
unsigned fr_cnt = 0;
|
||||
auto_vec<tree> fr_list;
|
||||
std::string str;
|
||||
|
||||
/* Old Format: const-int-expr-or-string-literal-list */
|
||||
if (cp_lexer_peek_token (parser->lexer)->type != CPP_OPEN_BRACE)
|
||||
while (true)
|
||||
{
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
if (cp_lexer_peek_token (parser->lexer)->type == CPP_STRING)
|
||||
{
|
||||
cp_expr cval = cp_parser_unevaluated_string_literal (parser);
|
||||
if (*cval == error_mark_node)
|
||||
return error_mark_node;
|
||||
if ((size_t) TREE_STRING_LENGTH (*cval)
|
||||
!= strlen (TREE_STRING_POINTER (*cval)) + 1)
|
||||
{
|
||||
error_at (cval.get_location (), "string literal must "
|
||||
"not contain %<\\0%>");
|
||||
return error_mark_node;
|
||||
}
|
||||
char c = omp_get_fr_id_from_name (TREE_STRING_POINTER (*cval));
|
||||
if (c == GOMP_INTEROP_IFR_UNKNOWN)
|
||||
warning_at (cval.get_location (), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qs",
|
||||
TREE_STRING_POINTER (*cval));
|
||||
str += c;
|
||||
}
|
||||
else
|
||||
{
|
||||
location_t loc = cp_lexer_peek_token (parser->lexer)->location;
|
||||
tree value = cp_parser_assignment_expression (parser);
|
||||
value = cp_fully_fold (value);
|
||||
if (!processing_template_decl)
|
||||
{
|
||||
if (TREE_CODE (value) != INTEGER_CST
|
||||
|| !tree_fits_shwi_p (value))
|
||||
{
|
||||
cp_parser_error (parser,
|
||||
"expected string literal or constant "
|
||||
"integer expression");
|
||||
return error_mark_node;
|
||||
}
|
||||
HOST_WIDE_INT n = tree_to_shwi (value);
|
||||
if (n < 1 || n > GOMP_INTEROP_IFR_LAST)
|
||||
{
|
||||
warning_at (EXPR_LOC_OR_LOC (value, loc), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qwd", n);
|
||||
n = GOMP_INTEROP_IFR_UNKNOWN;
|
||||
}
|
||||
str += (char) n;
|
||||
}
|
||||
else
|
||||
{
|
||||
str += (char) GOMP_INTEROP_IFR_UNKNOWN;
|
||||
for (unsigned n = fr_list.length (); n < fr_cnt; n++)
|
||||
fr_list.safe_push (NULL_TREE);
|
||||
if (EXPR_LOCATION (value) == UNKNOWN_LOCATION)
|
||||
value = build1_loc (loc, NOP_EXPR, TREE_TYPE (value), value);
|
||||
fr_list.safe_push (value);
|
||||
}
|
||||
}
|
||||
fr_cnt++;
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
str += '\0';
|
||||
if (cp_lexer_peek_token (parser->lexer)->type == CPP_COMMA)
|
||||
{
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
continue;
|
||||
}
|
||||
if (cp_lexer_peek_token (parser->lexer)->type != CPP_CLOSE_PAREN)
|
||||
{
|
||||
cp_parser_error (parser, "expected %<,%> or %<)%>");
|
||||
return error_mark_node;
|
||||
}
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
str += '\0';
|
||||
tree res = build_string (str.length (), str.data ());
|
||||
TREE_TYPE (res) = build_array_type_nelts (unsigned_char_type_node,
|
||||
str.length ());
|
||||
if (!fr_list.is_empty ())
|
||||
{
|
||||
tree t = make_tree_vec (fr_list.length ());
|
||||
for (unsigned i = 0; i < fr_list.length (); i++)
|
||||
TREE_VEC_ELT (t, i) = fr_list[i];
|
||||
res = tree_cons (res, t, NULL_TREE);
|
||||
}
|
||||
return res;
|
||||
}
|
||||
|
||||
/* New format. */
|
||||
std::string str2;
|
||||
while (true)
|
||||
{
|
||||
if (!cp_parser_require (parser, CPP_OPEN_BRACE, RT_OPEN_BRACE))
|
||||
return error_mark_node;
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
str2.clear ();
|
||||
bool has_fr = false;
|
||||
while (true)
|
||||
{
|
||||
cp_token *tok = cp_lexer_peek_token (parser->lexer);
|
||||
if (tok->type != CPP_NAME
|
||||
|| (strcmp("fr", IDENTIFIER_POINTER (tok->u.value)) != 0
|
||||
&& strcmp("attr", IDENTIFIER_POINTER (tok->u.value)) != 0))
|
||||
{
|
||||
cp_parser_error (parser, "expected %<fr%> or %<attr%> preference "
|
||||
"selector");
|
||||
return error_mark_node;
|
||||
}
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
bool is_fr = IDENTIFIER_POINTER (tok->u.value)[0] == 'f';
|
||||
if (is_fr && has_fr)
|
||||
{
|
||||
cp_parser_error (parser, "duplicated %<fr%> preference selector");
|
||||
return error_mark_node;
|
||||
}
|
||||
if (!cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN))
|
||||
return error_mark_node;
|
||||
while (true)
|
||||
{
|
||||
if (cp_lexer_peek_token (parser->lexer)->type == CPP_STRING)
|
||||
{
|
||||
cp_expr cval = cp_parser_unevaluated_string_literal (parser);
|
||||
if (*cval == error_mark_node)
|
||||
return error_mark_node;
|
||||
if ((size_t) TREE_STRING_LENGTH (*cval)
|
||||
!= strlen (TREE_STRING_POINTER (*cval)) + 1)
|
||||
{
|
||||
error_at (cval.get_location (), "string literal must "
|
||||
"not contain %<\\0%>");
|
||||
return error_mark_node;
|
||||
}
|
||||
if (!is_fr)
|
||||
{
|
||||
if (!startswith (TREE_STRING_POINTER (*cval), "ompx_"))
|
||||
{
|
||||
error_at (cval.get_location (),
|
||||
"%<attr%> string literal must start with "
|
||||
"%<ompx_%>");
|
||||
return error_mark_node;
|
||||
}
|
||||
if (strchr (TREE_STRING_POINTER (*cval), ','))
|
||||
{
|
||||
error_at (cval.get_location (),
|
||||
"%<attr%> string literal must not contain "
|
||||
"a comma");
|
||||
return error_mark_node;
|
||||
}
|
||||
str2 += TREE_STRING_POINTER (*cval);
|
||||
str2 += '\0';
|
||||
}
|
||||
else
|
||||
{
|
||||
if (*TREE_STRING_POINTER (*cval) == '\0')
|
||||
{
|
||||
cp_parser_error (parser,
|
||||
"non-empty string literal expected");
|
||||
return error_mark_node;
|
||||
}
|
||||
char c
|
||||
= omp_get_fr_id_from_name (TREE_STRING_POINTER (*cval));
|
||||
if (c == GOMP_INTEROP_IFR_UNKNOWN)
|
||||
warning_at (cval.get_location (), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qs",
|
||||
TREE_STRING_POINTER (*cval));
|
||||
str += c;
|
||||
has_fr = true;
|
||||
}
|
||||
}
|
||||
else if (!is_fr)
|
||||
{
|
||||
cp_parser_error (parser, "expected string literal");
|
||||
return error_mark_node;
|
||||
}
|
||||
else
|
||||
{
|
||||
location_t loc = cp_lexer_peek_token (parser->lexer)->location;
|
||||
tree value = cp_parser_assignment_expression (parser);
|
||||
value = cp_fully_fold (value);
|
||||
if (!processing_template_decl)
|
||||
{
|
||||
if (TREE_CODE (value) != INTEGER_CST
|
||||
|| !tree_fits_shwi_p (value))
|
||||
{
|
||||
cp_parser_error (parser,
|
||||
"expected string literal or "
|
||||
"constant integer expression");
|
||||
return error_mark_node;
|
||||
}
|
||||
HOST_WIDE_INT n = tree_to_shwi (value);
|
||||
if (n < 1 || n > GOMP_INTEROP_IFR_LAST)
|
||||
{
|
||||
warning_at (EXPR_LOC_OR_LOC (value, loc), OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qwd",
|
||||
n);
|
||||
n = GOMP_INTEROP_IFR_UNKNOWN;
|
||||
}
|
||||
str += (char) n;
|
||||
}
|
||||
else
|
||||
{
|
||||
str += (char) GOMP_INTEROP_IFR_UNKNOWN;
|
||||
for (unsigned n = fr_list.length (); n < fr_cnt; n++)
|
||||
fr_list.safe_push (NULL_TREE);
|
||||
if (EXPR_LOCATION (value) == UNKNOWN_LOCATION)
|
||||
value = build1_loc (loc, NOP_EXPR, TREE_TYPE (value),
|
||||
value);
|
||||
fr_list.safe_push (value);
|
||||
}
|
||||
has_fr = true;
|
||||
}
|
||||
if (!is_fr
|
||||
&& cp_lexer_peek_token (parser->lexer)->type == CPP_COMMA)
|
||||
{
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
continue;
|
||||
}
|
||||
if (cp_lexer_peek_token (parser->lexer)->type != CPP_CLOSE_PAREN)
|
||||
{
|
||||
cp_parser_error (parser,
|
||||
is_fr ? G_("expected %<)%>")
|
||||
: G_("expected %<)%> or %<,%>"));
|
||||
return error_mark_node;
|
||||
}
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
break;
|
||||
}
|
||||
if (cp_lexer_peek_token (parser->lexer)->type == CPP_COMMA)
|
||||
{
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
continue;
|
||||
}
|
||||
if (cp_lexer_peek_token (parser->lexer)->type != CPP_CLOSE_BRACE)
|
||||
{
|
||||
cp_parser_error (parser, "expected %<,%> or %<}%>");
|
||||
return error_mark_node;
|
||||
}
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
break;
|
||||
}
|
||||
fr_cnt++;
|
||||
str += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
str += str2;
|
||||
str += '\0';
|
||||
if (cp_lexer_peek_token (parser->lexer)->type == CPP_CLOSE_PAREN)
|
||||
break;
|
||||
if (cp_lexer_peek_token (parser->lexer)->type != CPP_COMMA)
|
||||
{
|
||||
cp_parser_error (parser, "expected %<)%> or %<,%>");
|
||||
return error_mark_node;
|
||||
}
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
}
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
str += '\0';
|
||||
tree res = build_string (str.length (), str.data ());
|
||||
TREE_TYPE (res) = build_array_type_nelts (unsigned_char_type_node,
|
||||
str.length ());
|
||||
if (!fr_list.is_empty ())
|
||||
{
|
||||
tree t = make_tree_vec (fr_list.length ());
|
||||
for (unsigned i = 0; i < fr_list.length (); i++)
|
||||
TREE_VEC_ELT (t, i) = fr_list[i];
|
||||
res = tree_cons (res, t, NULL_TREE);
|
||||
}
|
||||
return res;
|
||||
}
|
||||
|
||||
/* OpenMP 5.1:
|
||||
init ( [init-modifier-list : ] variable-list )
|
||||
|
||||
Modifiers:
|
||||
target
|
||||
targetsync
|
||||
prefer_type (preference-specification) */
|
||||
|
||||
static tree
|
||||
cp_parser_omp_clause_init (cp_parser *parser, tree list)
|
||||
{
|
||||
if (!cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN))
|
||||
return list;
|
||||
|
||||
unsigned pos = 0, raw_pos = 1;
|
||||
while (cp_lexer_peek_nth_token (parser->lexer, raw_pos)->type == CPP_NAME)
|
||||
{
|
||||
pos++; raw_pos++;
|
||||
if (cp_lexer_peek_nth_token (parser->lexer, raw_pos)->type
|
||||
== CPP_OPEN_PAREN)
|
||||
{
|
||||
unsigned n = cp_parser_skip_balanced_tokens (parser, raw_pos);
|
||||
if (n == raw_pos)
|
||||
{
|
||||
pos = 0;
|
||||
break;
|
||||
}
|
||||
raw_pos = n;
|
||||
}
|
||||
if (cp_lexer_peek_nth_token (parser->lexer, raw_pos)->type == CPP_COLON)
|
||||
break;
|
||||
if (cp_lexer_peek_nth_token (parser->lexer, raw_pos)->type != CPP_COMMA)
|
||||
{
|
||||
pos = 0;
|
||||
break;
|
||||
}
|
||||
pos++;
|
||||
raw_pos++;
|
||||
}
|
||||
|
||||
bool target = false;
|
||||
bool targetsync = false;
|
||||
tree prefer_type_tree = NULL_TREE;
|
||||
|
||||
for (unsigned pos2 = 0; pos2 < pos; ++pos2)
|
||||
{
|
||||
cp_token *tok = cp_lexer_peek_token (parser->lexer);
|
||||
if (tok->type == CPP_COMMA)
|
||||
{
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
continue;
|
||||
}
|
||||
const char *p = IDENTIFIER_POINTER (tok->u.value);
|
||||
if (strcmp ("targetsync", p) == 0)
|
||||
{
|
||||
if (targetsync)
|
||||
error_at (tok->location, "duplicate %<targetsync%> modifier");
|
||||
targetsync = true;
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
}
|
||||
else if (strcmp ("target", p) == 0)
|
||||
{
|
||||
if (target)
|
||||
error_at (tok->location, "duplicate %<target%> modifier");
|
||||
target = true;
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
}
|
||||
else if (strcmp ("prefer_type", p) == 0)
|
||||
{
|
||||
if (prefer_type_tree != NULL_TREE)
|
||||
error_at (tok->location, "duplicate %<prefer_type%> modifier");
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
prefer_type_tree = cp_parser_omp_modifier_prefer_type (parser);
|
||||
if (prefer_type_tree == error_mark_node)
|
||||
return error_mark_node;
|
||||
}
|
||||
else
|
||||
{
|
||||
cp_parser_error (parser, "%<init%> clause with modifier other than "
|
||||
"%<prefer_type%>, %<target%> or "
|
||||
"%<targetsync%>");
|
||||
cp_parser_skip_to_closing_parenthesis (parser, /*recovering=*/true,
|
||||
/*or_comma=*/false,
|
||||
/*consume_paren=*/true);
|
||||
return list;
|
||||
}
|
||||
}
|
||||
if (pos)
|
||||
{
|
||||
gcc_checking_assert (cp_lexer_peek_token (parser->lexer)->type
|
||||
== CPP_COLON);
|
||||
cp_lexer_consume_token (parser->lexer);
|
||||
}
|
||||
|
||||
tree nl = cp_parser_omp_var_list_no_open (parser, OMP_CLAUSE_INIT, list,
|
||||
NULL, false);
|
||||
for (tree c = nl; c != list; c = OMP_CLAUSE_CHAIN (c))
|
||||
{
|
||||
if (target)
|
||||
OMP_CLAUSE_INIT_TARGET (c) = 1;
|
||||
if (targetsync)
|
||||
OMP_CLAUSE_INIT_TARGETSYNC (c) = 1;
|
||||
if (prefer_type_tree)
|
||||
OMP_CLAUSE_INIT_PREFER_TYPE (c) = prefer_type_tree;
|
||||
}
|
||||
return nl;
|
||||
}
|
||||
|
||||
/* OpenACC:
|
||||
async [( int-expr )] */
|
||||
|
||||
@ -42896,6 +43315,27 @@ cp_parser_omp_all_clauses (cp_parser *parser, omp_clause_mask mask,
|
||||
token->location);
|
||||
c_name = "doacross";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_DESTROY:
|
||||
clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_DESTROY,
|
||||
clauses);
|
||||
c_name = "destroy";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_INIT:
|
||||
{
|
||||
/* prefer_type parsing fails often such that many follow-up errors
|
||||
are printed and recovery by cp_parser_skip_to_closing_parenthesis
|
||||
will might skip to EOF, leading to an ICE elsewhere. */
|
||||
tree nc = cp_parser_omp_clause_init (parser, clauses);
|
||||
if (nc == error_mark_node)
|
||||
goto saw_error;
|
||||
clauses = nc;
|
||||
}
|
||||
c_name = "init";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_USE:
|
||||
clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_USE, clauses);
|
||||
c_name = "use";
|
||||
break;
|
||||
case PRAGMA_OMP_CLAUSE_DETACH:
|
||||
clauses = cp_parser_omp_clause_detach (parser, clauses);
|
||||
c_name = "detach";
|
||||
@ -42997,8 +43437,9 @@ cp_parser_omp_all_clauses (cp_parser *parser, omp_clause_mask mask,
|
||||
{
|
||||
if ((mask & (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_UNIFORM)) != 0)
|
||||
return finish_omp_clauses (clauses, C_ORT_OMP_DECLARE_SIMD);
|
||||
else
|
||||
return finish_omp_clauses (clauses, C_ORT_OMP);
|
||||
if ((mask & (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_USE)) != 0)
|
||||
return finish_omp_clauses (clauses, C_ORT_OMP_INTEROP);
|
||||
return finish_omp_clauses (clauses, C_ORT_OMP);
|
||||
}
|
||||
return clauses;
|
||||
}
|
||||
@ -50711,6 +51152,30 @@ cp_parser_omp_declare (cp_parser *parser, cp_token *pragma_tok,
|
||||
return false;
|
||||
}
|
||||
|
||||
/* OpenMP 5.1:
|
||||
# pragma omp interop clauses[opt] new-line */
|
||||
|
||||
#define OMP_INTEROP_CLAUSE_MASK \
|
||||
( (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DEPEND) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DESTROY) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DEVICE) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_INIT) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_NOWAIT) \
|
||||
| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_USE))
|
||||
|
||||
static void
|
||||
cp_parser_omp_interop (cp_parser *parser, cp_token *pragma_tok)
|
||||
{
|
||||
location_t loc = cp_lexer_peek_token (parser->lexer)->location;
|
||||
tree clauses = cp_parser_omp_all_clauses (parser, OMP_INTEROP_CLAUSE_MASK,
|
||||
"#pragma omp interop", pragma_tok);
|
||||
tree stmt = make_node (OMP_INTEROP);
|
||||
TREE_TYPE (stmt) = void_type_node;
|
||||
OMP_INTEROP_CLAUSES (stmt) = clauses;
|
||||
SET_EXPR_LOCATION (stmt, loc);
|
||||
add_stmt (stmt);
|
||||
}
|
||||
|
||||
/* OpenMP 5.0
|
||||
#pragma omp requires clauses[optseq] new-line */
|
||||
|
||||
@ -51990,6 +52455,22 @@ cp_parser_pragma (cp_parser *parser, enum pragma_context context, bool *if_p)
|
||||
}
|
||||
break;
|
||||
|
||||
case PRAGMA_OMP_INTEROP:
|
||||
switch (context)
|
||||
{
|
||||
case pragma_compound:
|
||||
cp_parser_omp_interop (parser, pragma_tok);
|
||||
return false;
|
||||
case pragma_stmt:
|
||||
error_at (pragma_tok->location, "%<#pragma %s%> may only be "
|
||||
"used in compound statements", "omp interop");
|
||||
ret = true;
|
||||
break;
|
||||
default:
|
||||
goto bad_stmt;
|
||||
}
|
||||
break;
|
||||
|
||||
case PRAGMA_OMP_TASKWAIT:
|
||||
switch (context)
|
||||
{
|
||||
|
||||
38
gcc/cp/pt.cc
38
gcc/cp/pt.cc
@ -17866,6 +17866,36 @@ tsubst_omp_clauses (tree clauses, enum c_omp_region_type ort,
|
||||
= tsubst_stmt (OMP_CLAUSE_LINEAR_STEP (oc), args,
|
||||
complain, in_decl);
|
||||
break;
|
||||
case OMP_CLAUSE_INIT:
|
||||
if (ort == C_ORT_OMP_INTEROP
|
||||
&& OMP_CLAUSE_INIT_PREFER_TYPE (nc)
|
||||
&& TREE_CODE (OMP_CLAUSE_INIT_PREFER_TYPE (nc)) == TREE_LIST
|
||||
&& (OMP_CLAUSE_CHAIN (nc) == NULL_TREE
|
||||
|| OMP_CLAUSE_CODE (OMP_CLAUSE_CHAIN (nc) ) != OMP_CLAUSE_INIT
|
||||
|| (OMP_CLAUSE_INIT_PREFER_TYPE (nc)
|
||||
!= OMP_CLAUSE_INIT_PREFER_TYPE (OMP_CLAUSE_CHAIN (nc) ))))
|
||||
{
|
||||
tree pref_list = OMP_CLAUSE_INIT_PREFER_TYPE (nc);
|
||||
tree fr_list = TREE_VALUE (pref_list);
|
||||
int len = TREE_VEC_LENGTH (fr_list);
|
||||
for (int i = 0; i < len; i++)
|
||||
{
|
||||
tree *fr_expr = &TREE_VEC_ELT (fr_list, i);
|
||||
/* Preserve NOP_EXPR to have a location. */
|
||||
if (*fr_expr && TREE_CODE (*fr_expr) == NOP_EXPR)
|
||||
TREE_OPERAND (*fr_expr, 0)
|
||||
= tsubst_expr (TREE_OPERAND (*fr_expr, 0), args, complain,
|
||||
in_decl);
|
||||
else
|
||||
*fr_expr = tsubst_expr (*fr_expr, args, complain, in_decl);
|
||||
}
|
||||
}
|
||||
/* FALLTHRU */
|
||||
case OMP_CLAUSE_DESTROY:
|
||||
case OMP_CLAUSE_USE:
|
||||
OMP_CLAUSE_OPERAND (nc, 0)
|
||||
= tsubst_stmt (OMP_CLAUSE_OPERAND (oc, 0), args, complain, in_decl);
|
||||
break;
|
||||
case OMP_CLAUSE_NOWAIT:
|
||||
case OMP_CLAUSE_DEFAULT:
|
||||
case OMP_CLAUSE_UNTIED:
|
||||
@ -19612,6 +19642,14 @@ tsubst_stmt (tree t, tree args, tsubst_flags_t complain, tree in_decl)
|
||||
}
|
||||
break;
|
||||
|
||||
case OMP_INTEROP:
|
||||
tmp = tsubst_omp_clauses (OMP_INTEROP_CLAUSES (t), C_ORT_OMP_INTEROP,
|
||||
args, complain, in_decl);
|
||||
t = copy_node (t);
|
||||
OMP_INTEROP_CLAUSES (t) = tmp;
|
||||
add_stmt (t);
|
||||
break;
|
||||
|
||||
case MUST_NOT_THROW_EXPR:
|
||||
{
|
||||
tree op0 = RECUR (TREE_OPERAND (t, 0));
|
||||
|
||||
@ -7208,6 +7208,77 @@ cp_oacc_check_attachments (tree c)
|
||||
return false;
|
||||
}
|
||||
|
||||
/* Update OMP_CLAUSE_INIT_PREFER_TYPE in case template substitution
|
||||
happened. */
|
||||
|
||||
static void
|
||||
cp_omp_init_prefer_type_update (tree c)
|
||||
{
|
||||
if (processing_template_decl
|
||||
|| OMP_CLAUSE_INIT_PREFER_TYPE (c) == NULL_TREE
|
||||
|| TREE_CODE (OMP_CLAUSE_INIT_PREFER_TYPE (c)) != TREE_LIST)
|
||||
return;
|
||||
|
||||
tree t = TREE_PURPOSE (OMP_CLAUSE_INIT_PREFER_TYPE (c));
|
||||
char *str = const_cast<char *> (TREE_STRING_POINTER (t));
|
||||
tree fr_list = TREE_VALUE (OMP_CLAUSE_INIT_PREFER_TYPE (c));
|
||||
int len = TREE_VEC_LENGTH (fr_list);
|
||||
int cnt = 0;
|
||||
|
||||
while (str[0] == (char) GOMP_INTEROP_IFR_SEPARATOR)
|
||||
{
|
||||
str++;
|
||||
if (str[0] == (char) GOMP_INTEROP_IFR_UNKNOWN)
|
||||
{
|
||||
/* Assume a no or a single 'fr'. */
|
||||
gcc_checking_assert (str[1] == (char) GOMP_INTEROP_IFR_SEPARATOR);
|
||||
location_t loc = UNKNOWN_LOCATION;
|
||||
tree value = TREE_VEC_ELT (fr_list, cnt);
|
||||
if (value != NULL_TREE && value != error_mark_node)
|
||||
{
|
||||
loc = EXPR_LOCATION (value);
|
||||
if (value && TREE_CODE (value) == NOP_EXPR)
|
||||
value = TREE_OPERAND (value, 0);
|
||||
value = cp_fully_fold (value);
|
||||
}
|
||||
if (value != NULL_TREE && value != error_mark_node)
|
||||
{
|
||||
if (TREE_CODE (value) != INTEGER_CST
|
||||
|| !tree_fits_shwi_p (value))
|
||||
error_at (loc,
|
||||
"expected string literal or "
|
||||
"constant integer expression instead of %qE", value); // FIXME of 'qE' and no 'loc'?
|
||||
else
|
||||
{
|
||||
HOST_WIDE_INT n = tree_to_shwi (value);
|
||||
if (n < 1 || n > GOMP_INTEROP_IFR_LAST)
|
||||
{
|
||||
warning_at (loc, OPT_Wopenmp,
|
||||
"unknown foreign runtime identifier %qwd", n);
|
||||
n = GOMP_INTEROP_IFR_UNKNOWN;
|
||||
}
|
||||
str[0] = (char) n;
|
||||
}
|
||||
}
|
||||
str++;
|
||||
}
|
||||
else if (str[0] != (char) GOMP_INTEROP_IFR_SEPARATOR)
|
||||
{
|
||||
/* Assume a no or a single 'fr'. */
|
||||
gcc_checking_assert (str[1] == (char) GOMP_INTEROP_IFR_SEPARATOR);
|
||||
str++;
|
||||
}
|
||||
str++;
|
||||
while (str[0] != '\0')
|
||||
str += strlen (str) + 1;
|
||||
str++;
|
||||
cnt++;
|
||||
if (cnt >= len)
|
||||
break;
|
||||
}
|
||||
OMP_CLAUSE_INIT_PREFER_TYPE (c) = t;
|
||||
}
|
||||
|
||||
/* For all elements of CLAUSES, validate them vs OpenMP constraints.
|
||||
Remove any elements from the list that are invalid. */
|
||||
|
||||
@ -7239,6 +7310,10 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
bool target_in_reduction_seen = false;
|
||||
bool num_tasks_seen = false;
|
||||
bool partial_seen = false;
|
||||
bool init_seen = false;
|
||||
bool init_use_destroy_seen = false;
|
||||
tree init_no_targetsync_clause = NULL_TREE;
|
||||
tree depend_clause = NULL_TREE;
|
||||
|
||||
bitmap_obstack_initialize (NULL);
|
||||
bitmap_initialize (&generic_head, &bitmap_default_obstack);
|
||||
@ -8347,6 +8422,8 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
}
|
||||
gcc_unreachable ();
|
||||
case OMP_CLAUSE_DEPEND:
|
||||
depend_clause = c;
|
||||
/* FALLTHRU */
|
||||
case OMP_CLAUSE_AFFINITY:
|
||||
t = OMP_CLAUSE_DECL (c);
|
||||
if (TREE_CODE (t) == TREE_LIST
|
||||
@ -9356,7 +9433,37 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
|
||||
OMP_CLAUSE_PARTIAL_EXPR (c) = t;
|
||||
break;
|
||||
|
||||
case OMP_CLAUSE_INIT:
|
||||
init_seen = true;
|
||||
cp_omp_init_prefer_type_update (c);
|
||||
if (!OMP_CLAUSE_INIT_TARGETSYNC (c))
|
||||
init_no_targetsync_clause = c;
|
||||
/* FALLTHRU */
|
||||
case OMP_CLAUSE_DESTROY:
|
||||
case OMP_CLAUSE_USE:
|
||||
init_use_destroy_seen = true;
|
||||
t = OMP_CLAUSE_DECL (c);
|
||||
if (bitmap_bit_p (&generic_head, DECL_UID (t)))
|
||||
{
|
||||
error_at (OMP_CLAUSE_LOCATION (c),
|
||||
"%qD appears more than once in action clauses", t);
|
||||
remove = true;
|
||||
}
|
||||
if (!processing_template_decl)
|
||||
{
|
||||
if (/* ort == C_ORT_OMP_INTEROP [uncomment for depobj init] */
|
||||
!c_omp_interop_t_p (TREE_TYPE (OMP_CLAUSE_DECL (c))))
|
||||
error_at (OMP_CLAUSE_LOCATION (c),
|
||||
"%qD must be of %<omp_interop_t%>",
|
||||
OMP_CLAUSE_DECL (c));
|
||||
else if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_USE
|
||||
&& TREE_READONLY (OMP_CLAUSE_DECL (c)))
|
||||
error_at (OMP_CLAUSE_LOCATION (c),
|
||||
"%qD shall not be const", OMP_CLAUSE_DECL (c));
|
||||
}
|
||||
bitmap_set_bit (&generic_head, DECL_UID (t));
|
||||
pc = &OMP_CLAUSE_CHAIN (c);
|
||||
break;
|
||||
default:
|
||||
gcc_unreachable ();
|
||||
}
|
||||
@ -9794,6 +9901,19 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
|
||||
pc = &OMP_CLAUSE_CHAIN (c);
|
||||
}
|
||||
|
||||
if (ort == C_ORT_OMP_INTEROP
|
||||
&& depend_clause
|
||||
&& (!init_use_destroy_seen
|
||||
|| (init_seen && init_no_targetsync_clause)))
|
||||
{
|
||||
error_at (OMP_CLAUSE_LOCATION (depend_clause),
|
||||
"%<depend%> clause requires action clauses with "
|
||||
"%<targetsync%> interop-type");
|
||||
if (init_no_targetsync_clause)
|
||||
inform (OMP_CLAUSE_LOCATION (init_no_targetsync_clause),
|
||||
"%<init%> clause lacks the %<targetsync%> modifier");
|
||||
}
|
||||
|
||||
bitmap_obstack_release (NULL);
|
||||
return clauses;
|
||||
}
|
||||
|
||||
@ -1544,63 +1544,42 @@ show_omp_namelist (int list_type, gfc_omp_namelist *n)
|
||||
fputs ("target,", dumpfile);
|
||||
if (n->u.init.targetsync)
|
||||
fputs ("targetsync,", dumpfile);
|
||||
if (n->u2.init_interop_fr)
|
||||
if (n->u2.init_interop)
|
||||
{
|
||||
char *attr_str = n->u.init.attr;
|
||||
int idx = 0;
|
||||
int fr_id;
|
||||
char *str = n->u2.init_interop;
|
||||
fputs ("prefer_type(", dumpfile);
|
||||
do
|
||||
while (str[0] == (char) GOMP_INTEROP_IFR_SEPARATOR)
|
||||
{
|
||||
fr_id = n->u2.init_interop_fr[idx];
|
||||
bool has_fr = false;
|
||||
fputc ('{', dumpfile);
|
||||
if (fr_id != GOMP_INTEROP_IFR_NONE)
|
||||
str++;
|
||||
while (str[0] != (char) GOMP_INTEROP_IFR_SEPARATOR)
|
||||
{
|
||||
fputs ("fr(", dumpfile);
|
||||
do
|
||||
{
|
||||
const char *fr_str = omp_get_name_from_fr_id (fr_id);
|
||||
if (fr_str)
|
||||
fprintf (dumpfile, "\"%s\"", fr_str);
|
||||
else
|
||||
fprintf (dumpfile, "%d", fr_id);
|
||||
fr_id = n->u2.init_interop_fr[++idx];
|
||||
if (fr_id != GOMP_INTEROP_IFR_SEPARATOR)
|
||||
fputc (',', dumpfile);
|
||||
}
|
||||
while (fr_id != GOMP_INTEROP_IFR_SEPARATOR);
|
||||
fputc (')', dumpfile);
|
||||
if (attr_str && (attr_str[0] != ' ' || attr_str[1] != '\0'))
|
||||
if (has_fr)
|
||||
fputc (',', dumpfile);
|
||||
has_fr = true;
|
||||
fputs ("fr(\"", dumpfile);
|
||||
fputs (omp_get_name_from_fr_id (str[0]), dumpfile);
|
||||
fputs ("\")", dumpfile);
|
||||
str++;
|
||||
}
|
||||
else
|
||||
fr_id = n->u2.init_interop_fr[++idx];
|
||||
if (attr_str && attr_str[0] == ' ' && attr_str[1] == '\0')
|
||||
attr_str += 2;
|
||||
else if (attr_str)
|
||||
str++;
|
||||
if (has_fr && str[0] != '\0')
|
||||
fputc (',', dumpfile);
|
||||
while (str[0] != '\0')
|
||||
{
|
||||
fputs ("attr(\"", dumpfile);
|
||||
do
|
||||
{
|
||||
fputs ((char *) attr_str, dumpfile);
|
||||
fputc ('"', dumpfile);
|
||||
attr_str += strlen (attr_str) + 1;
|
||||
if (attr_str[0] == '\0')
|
||||
break;
|
||||
fputs (",\"", dumpfile);
|
||||
}
|
||||
while (true);
|
||||
fputc (')', dumpfile);
|
||||
fputs (str, dumpfile);
|
||||
fputs ("\")", dumpfile);
|
||||
str += strlen (str) + 1;
|
||||
if (str[0] != '\0')
|
||||
fputc (',', dumpfile);
|
||||
}
|
||||
str++;
|
||||
fputc ('}', dumpfile);
|
||||
fr_id = n->u2.init_interop_fr[++idx];
|
||||
if (fr_id == GOMP_INTEROP_IFR_SEPARATOR)
|
||||
break;
|
||||
fputc (',', dumpfile);
|
||||
if (attr_str)
|
||||
++attr_str;
|
||||
if (str[0] != '\0')
|
||||
fputs (", ", dumpfile);
|
||||
}
|
||||
while (true);
|
||||
fputc (')', dumpfile);
|
||||
}
|
||||
fputc (':', dumpfile);
|
||||
|
||||
@ -1403,7 +1403,6 @@ typedef struct gfc_omp_namelist
|
||||
bool present_modifier;
|
||||
struct
|
||||
{
|
||||
char *attr;
|
||||
int len;
|
||||
bool target;
|
||||
bool targetsync;
|
||||
@ -1416,7 +1415,7 @@ typedef struct gfc_omp_namelist
|
||||
gfc_expr *allocator;
|
||||
struct gfc_symbol *traits_sym;
|
||||
struct gfc_omp_namelist *duplicate_of;
|
||||
char *init_interop_fr;
|
||||
char *init_interop;
|
||||
} u2;
|
||||
struct gfc_omp_namelist *next;
|
||||
locus where;
|
||||
|
||||
@ -5608,7 +5608,7 @@ gfc_free_omp_namelist (gfc_omp_namelist *name, bool free_ns,
|
||||
{
|
||||
gfc_omp_namelist *n;
|
||||
gfc_expr *last_allocator = NULL;
|
||||
char *last_init_attr = NULL;
|
||||
char *last_init_interop = NULL;
|
||||
|
||||
for (; name; name = n)
|
||||
{
|
||||
@ -5632,11 +5632,10 @@ gfc_free_omp_namelist (gfc_omp_namelist *name, bool free_ns,
|
||||
{ } /* name->u2.traits_sym: shall not call gfc_free_symbol here. */
|
||||
else if (free_init)
|
||||
{
|
||||
if (name->u.init.attr != last_init_attr)
|
||||
if (name->u2.init_interop != last_init_interop)
|
||||
{
|
||||
last_init_attr = name->u.init.attr;
|
||||
free (name->u.init.attr);
|
||||
free (name->u2.init_interop_fr);
|
||||
last_init_interop = name->u2.init_interop;
|
||||
free (name->u2.init_interop);
|
||||
}
|
||||
}
|
||||
else if (name->u2.udr)
|
||||
|
||||
@ -1835,45 +1835,42 @@ error:
|
||||
prefer_type ( <const-int-expr|string literal> [, ...]
|
||||
or
|
||||
prefer_type ( '{' <fr(...) | attr (...)>, ...] '}' [, '{' ... '}' ] )
|
||||
where 'fr' takes an integer named constant or a string literal
|
||||
and 'attr takes a string literal, starting with 'ompx_')
|
||||
where 'fr' takes a constant expression or a string literal
|
||||
and 'attr takes a list of string literals, starting with 'ompx_')
|
||||
|
||||
For the foreign runtime identifiers, string values are converted to
|
||||
their integer value; unknown string or integer values are set to 0.
|
||||
their integer value; unknown string or integer values are set to
|
||||
GOMP_INTEROP_IFR_KNOWN.
|
||||
|
||||
For the simple syntax, pref_int_array contains alternatingly the
|
||||
fr_id integer value and GOMP_INTEROP_IFR_SEPARATOR followed by a
|
||||
GOMP_INTEROP_IFR_SEPARATOR as last item.
|
||||
For the complex syntax, it contains the values associated with a
|
||||
'fr(...)' followed by GOMP_INTEROP_IFR_SEPARATOR. If there is no
|
||||
'fr' in a curly-brace block, it is GOMP_INTEROP_IFR_NONE followed
|
||||
by GOMP_INTEROP_IFR_SEPARATOR. An additional GOMP_INTEROP_IFR_SEPARATOR
|
||||
at the end terminates the array.
|
||||
Data format:
|
||||
For the foreign runtime identifiers, string values are converted to
|
||||
their integer value; unknown string or integer values are set to 0.
|
||||
|
||||
For attributes, if the simply syntax is used, it is NULL - likewise if no
|
||||
'attr' appears. For the complex syntax it is: For reach curly-brace block,
|
||||
it is \0\0 is no attr appears and otherwise a concatenation (including
|
||||
the \0) of all 'attr' strings followed by a tailing '\0'. At the end,
|
||||
another '\0' follows. */
|
||||
Each item (a) GOMP_INTEROP_IFR_SEPARATOR
|
||||
(b) for any 'fr', its integer value.
|
||||
Note: Spec only permits 1 'fr' entry (6.0; changed after TR13)
|
||||
(c) GOMP_INTEROP_IFR_SEPARATOR
|
||||
(d) list of \0-terminated non-empty strings for 'attr'
|
||||
(e) '\0'
|
||||
Tailing '\0'. */
|
||||
|
||||
static match
|
||||
gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_len)
|
||||
gfc_match_omp_prefer_type (char **type_str, int *type_str_len)
|
||||
{
|
||||
gfc_expr *e;
|
||||
int cnt_brace_grp = 0;
|
||||
std::vector<char> int_list;
|
||||
std::string attr_string;
|
||||
std::string type_string, attr_string;
|
||||
/* New syntax. */
|
||||
if (gfc_peek_ascii_char () == '{')
|
||||
do
|
||||
{
|
||||
attr_string.clear ();
|
||||
type_string += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
if (gfc_match ("{ ") != MATCH_YES)
|
||||
{
|
||||
gfc_error ("Expected %<{%> at %C");
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
bool fr_found = false;
|
||||
bool attr_found = false;
|
||||
do
|
||||
{
|
||||
if (gfc_match ("fr ( ") == MATCH_YES)
|
||||
@ -1887,24 +1884,27 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
fr_found = true;
|
||||
do
|
||||
{
|
||||
if (gfc_match_expr (&e) != MATCH_YES)
|
||||
return MATCH_ERROR;
|
||||
if (e->expr_type != EXPR_CONSTANT
|
||||
|| e->ref != NULL
|
||||
bool found_literal = false;
|
||||
match m = MATCH_YES;
|
||||
if (gfc_match_literal_constant (&e, false) == MATCH_YES)
|
||||
found_literal = true;
|
||||
else
|
||||
m = gfc_match_expr (&e);
|
||||
if (m != MATCH_YES
|
||||
|| !gfc_resolve_expr (e)
|
||||
|| e->rank != 0
|
||||
|| e->expr_type != EXPR_CONSTANT
|
||||
|| (e->ts.type != BT_INTEGER
|
||||
&& e->ts.type != BT_CHARACTER)
|
||||
&& (!found_literal || e->ts.type != BT_CHARACTER))
|
||||
|| (e->ts.type == BT_INTEGER
|
||||
&& (!e->symtree
|
||||
|| e->symtree->n.sym->attr.flavor != FL_PARAMETER
|
||||
|| !mpz_fits_sint_p (e->value.integer)))
|
||||
&& !mpz_fits_sint_p (e->value.integer))
|
||||
|| (e->ts.type == BT_CHARACTER
|
||||
&& (e->ts.kind != gfc_default_character_kind
|
||||
|| e->value.character.length == 0)))
|
||||
|| e->value.character.length == 0)))
|
||||
{
|
||||
gfc_error ("Expected scalar integer parameter or "
|
||||
"non-empty default-kind character literal "
|
||||
"at %L", &e->where);
|
||||
gfc_error ("Expected constant scalar integer expression"
|
||||
" or non-empty default-kind character "
|
||||
"literal at %L", &e->where);
|
||||
gfc_free_expr (e);
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
@ -1915,10 +1915,11 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
val = mpz_get_si (e->value.integer);
|
||||
if (val < 1 || val > GOMP_INTEROP_IFR_LAST)
|
||||
{
|
||||
gfc_warning (OPT_Wopenmp,
|
||||
"Unknown foreign runtime identifier "
|
||||
"%qd at %L", val, &e->where);
|
||||
val = 0;
|
||||
gfc_warning_now (OPT_Wopenmp,
|
||||
"Unknown foreign runtime "
|
||||
"identifier %qd at %L",
|
||||
val, &e->where);
|
||||
val = GOMP_INTEROP_IFR_UNKNOWN;
|
||||
}
|
||||
}
|
||||
else
|
||||
@ -1934,40 +1935,30 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
val = omp_get_fr_id_from_name (str);
|
||||
if (val == 0)
|
||||
gfc_warning (OPT_Wopenmp,
|
||||
"Unknown foreign runtime identifier %qs "
|
||||
"at %L", str, &e->where);
|
||||
if (val == GOMP_INTEROP_IFR_UNKNOWN)
|
||||
gfc_warning_now (OPT_Wopenmp,
|
||||
"Unknown foreign runtime identifier "
|
||||
"%qs at %L", str, &e->where);
|
||||
}
|
||||
int_list.push_back (val);
|
||||
if (gfc_match (", ") == MATCH_YES)
|
||||
continue;
|
||||
|
||||
type_string += (char) val;
|
||||
if (gfc_match (") ") == MATCH_YES)
|
||||
break;
|
||||
gfc_error ("Expected %<,%> or %<)%> at %C");
|
||||
gfc_error ("Expected %<)%> at %C");
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
while (true);
|
||||
}
|
||||
else if (gfc_match ("attr ( ") == MATCH_YES)
|
||||
{
|
||||
attr_found = true;
|
||||
if (attr_string.empty ())
|
||||
for (int i = 0; i < cnt_brace_grp; ++i)
|
||||
{
|
||||
/* Add dummy elements for previous curly-brace blocks. */
|
||||
attr_string += ' ';
|
||||
attr_string += '\0';
|
||||
attr_string += '\0';
|
||||
}
|
||||
do
|
||||
{
|
||||
if (gfc_match_expr (&e) != MATCH_YES)
|
||||
return MATCH_ERROR;
|
||||
if (e->expr_type != EXPR_CONSTANT
|
||||
if (gfc_match_literal_constant (&e, false) != MATCH_YES
|
||||
|| !gfc_resolve_expr (e)
|
||||
|| e->expr_type != EXPR_CONSTANT
|
||||
|| e->rank != 0
|
||||
|| e->ts.type != BT_CHARACTER
|
||||
|| e->ts.kind != gfc_default_character_kind)
|
||||
|| e->ts.kind != gfc_default_character_kind)
|
||||
{
|
||||
gfc_error ("Expected default-kind character literal "
|
||||
"at %L", &e->where);
|
||||
@ -2016,21 +2007,9 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
while (true);
|
||||
++cnt_brace_grp;
|
||||
if (!fr_found)
|
||||
int_list.push_back (GOMP_INTEROP_IFR_NONE);
|
||||
int_list.push_back (GOMP_INTEROP_IFR_SEPARATOR);
|
||||
if (!attr_string.empty ())
|
||||
{
|
||||
if (!attr_found)
|
||||
{
|
||||
/* Dummy entry. */
|
||||
attr_string += ' ';
|
||||
attr_string += '\0';
|
||||
}
|
||||
attr_string += '\0';
|
||||
}
|
||||
|
||||
type_string += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
type_string += attr_string;
|
||||
type_string += '\0';
|
||||
if (gfc_match (", ") == MATCH_YES)
|
||||
continue;
|
||||
if (gfc_match (") ") == MATCH_YES)
|
||||
@ -2042,12 +2021,19 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
else
|
||||
do
|
||||
{
|
||||
if (gfc_match_expr (&e) != MATCH_YES)
|
||||
return MATCH_ERROR;
|
||||
if (!gfc_resolve_expr (e)
|
||||
type_string += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
bool found_literal = false;
|
||||
match m = MATCH_YES;
|
||||
if (gfc_match_literal_constant (&e, false) == MATCH_YES)
|
||||
found_literal = true;
|
||||
else
|
||||
m = gfc_match_expr (&e);
|
||||
if (m != MATCH_YES
|
||||
|| !gfc_resolve_expr (e)
|
||||
|| e->rank != 0
|
||||
|| e->expr_type != EXPR_CONSTANT
|
||||
|| (e->ts.type != BT_INTEGER && e->ts.type != BT_CHARACTER)
|
||||
|| (e->ts.type != BT_INTEGER
|
||||
&& (!found_literal || e->ts.type != BT_CHARACTER))
|
||||
|| (e->ts.type == BT_INTEGER
|
||||
&& !mpz_fits_sint_p (e->value.integer))
|
||||
|| (e->ts.type == BT_CHARACTER
|
||||
@ -2066,9 +2052,9 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
val = mpz_get_si (e->value.integer);
|
||||
if (val < 1 || val > GOMP_INTEROP_IFR_LAST)
|
||||
{
|
||||
gfc_warning (OPT_Wopenmp,
|
||||
"Unknown foreign runtime identifier %qd at %L",
|
||||
val, &e->where);
|
||||
gfc_warning_now (OPT_Wopenmp,
|
||||
"Unknown foreign runtime identifier %qd at %L",
|
||||
val, &e->where);
|
||||
val = 0;
|
||||
}
|
||||
}
|
||||
@ -2084,13 +2070,14 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
val = omp_get_fr_id_from_name (str);
|
||||
if (val == 0)
|
||||
gfc_warning (OPT_Wopenmp,
|
||||
"Unknown foreign runtime identifier %qs at %L",
|
||||
str, &e->where);
|
||||
if (val == GOMP_INTEROP_IFR_UNKNOWN)
|
||||
gfc_warning_now (OPT_Wopenmp,
|
||||
"Unknown foreign runtime identifier %qs at %L",
|
||||
str, &e->where);
|
||||
}
|
||||
int_list.push_back (val);
|
||||
int_list.push_back (GOMP_INTEROP_IFR_SEPARATOR);
|
||||
type_string += (char) val;
|
||||
type_string += (char) GOMP_INTEROP_IFR_SEPARATOR;
|
||||
type_string += '\0';
|
||||
gfc_free_expr (e);
|
||||
if (gfc_match (", ") == MATCH_YES)
|
||||
continue;
|
||||
@ -2100,17 +2087,10 @@ gfc_match_omp_prefer_type (char **fr_int_array, char **attr_str, int *attr_str_l
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
while (true);
|
||||
int_list.push_back (GOMP_INTEROP_IFR_SEPARATOR);
|
||||
*fr_int_array = XNEWVEC (char, int_list.size ());
|
||||
memcpy (*fr_int_array, int_list.data (), sizeof (char) * int_list.size ());
|
||||
|
||||
if (!attr_string.empty ())
|
||||
{
|
||||
attr_string += '\0';
|
||||
*attr_str_len = attr_string.length();
|
||||
*attr_str = XNEWVEC (char, attr_string.length ());
|
||||
memcpy (*attr_str, attr_string.data (), attr_string.length ());
|
||||
}
|
||||
type_string += '\0';
|
||||
*type_str_len = type_string.length();
|
||||
*type_str = XNEWVEC (char, type_string.length ());
|
||||
memcpy (*type_str, type_string.data (), type_string.length ());
|
||||
return MATCH_YES;
|
||||
}
|
||||
|
||||
@ -2122,21 +2102,19 @@ static match
|
||||
gfc_match_omp_init (gfc_omp_namelist **list)
|
||||
{
|
||||
bool target = false, targetsync = false;
|
||||
char *fr_int_array = NULL;
|
||||
char *attr_str = NULL;
|
||||
int attr_str_len = 0;
|
||||
char *type_str = NULL;
|
||||
int type_str_len = 0;
|
||||
match m;
|
||||
locus old_loc = gfc_current_locus;
|
||||
do {
|
||||
if (gfc_match ("prefer_type ( ") == MATCH_YES)
|
||||
{
|
||||
if (fr_int_array)
|
||||
if (type_str)
|
||||
{
|
||||
gfc_error ("Duplicate %<prefer_type%> modifier at %C");
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
m = gfc_match_omp_prefer_type (&fr_int_array, &attr_str,
|
||||
&attr_str_len);
|
||||
m = gfc_match_omp_prefer_type (&type_str, &type_str_len);
|
||||
if (m != MATCH_YES)
|
||||
return m;
|
||||
if (gfc_match (", ") == MATCH_YES)
|
||||
@ -2148,16 +2126,21 @@ gfc_match_omp_init (gfc_omp_namelist **list)
|
||||
}
|
||||
if (gfc_match ("targetsync ") == MATCH_YES)
|
||||
{
|
||||
if (targetsync)
|
||||
{
|
||||
/* Avoid the word 'modifier' as it could be also be no clauses and
|
||||
twice a variable named 'targetsync', which is also invalid. */
|
||||
gfc_error ("Duplicate %<targetsync%> at %C");
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
targetsync = true;
|
||||
if (gfc_match (", ") == MATCH_YES)
|
||||
continue;
|
||||
if (gfc_match (": ") == MATCH_YES)
|
||||
break;
|
||||
gfc_char_t c = gfc_peek_char ();
|
||||
if (!fr_int_array
|
||||
&& (c == ')'
|
||||
|| (gfc_current_form != FORM_FREE
|
||||
&& (c == '_' || ISALPHA (c)))))
|
||||
if (!type_str && (c == ')' || (gfc_current_form != FORM_FREE
|
||||
&& (c == '_' || ISALPHA (c)))))
|
||||
{
|
||||
gfc_current_locus = old_loc;
|
||||
break;
|
||||
@ -2167,16 +2150,19 @@ gfc_match_omp_init (gfc_omp_namelist **list)
|
||||
}
|
||||
if (gfc_match ("target ") == MATCH_YES)
|
||||
{
|
||||
if (target)
|
||||
{
|
||||
gfc_error ("Duplicate %<target%> at %C");
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
target = true;
|
||||
if (gfc_match (", ") == MATCH_YES)
|
||||
continue;
|
||||
if (gfc_match (": ") == MATCH_YES)
|
||||
break;
|
||||
gfc_char_t c = gfc_peek_char ();
|
||||
if (!fr_int_array
|
||||
&& (c == ')'
|
||||
|| (gfc_current_form != FORM_FREE
|
||||
&& (c == '_' || ISALPHA (c)))))
|
||||
if (!type_str && (c == ')' || (gfc_current_form != FORM_FREE
|
||||
&& (c == '_' || ISALPHA (c)))))
|
||||
{
|
||||
gfc_current_locus = old_loc;
|
||||
break;
|
||||
@ -2184,7 +2170,7 @@ gfc_match_omp_init (gfc_omp_namelist **list)
|
||||
gfc_error ("Expected %<,%> or %<:%> at %C");
|
||||
return MATCH_ERROR;
|
||||
}
|
||||
if (fr_int_array)
|
||||
if (type_str)
|
||||
{
|
||||
gfc_error ("Expected %<target%> or %<targetsync%> at %C");
|
||||
return MATCH_ERROR;
|
||||
@ -2201,9 +2187,8 @@ gfc_match_omp_init (gfc_omp_namelist **list)
|
||||
{
|
||||
n->u.init.target = target;
|
||||
n->u.init.targetsync = targetsync;
|
||||
n->u.init.attr = attr_str;
|
||||
n->u.init.len = attr_str_len;
|
||||
n->u2.init_interop_fr = fr_int_array;
|
||||
n->u.init.len = type_str_len;
|
||||
n->u2.init_interop = type_str;
|
||||
}
|
||||
return MATCH_YES;
|
||||
}
|
||||
@ -8459,6 +8444,21 @@ resolve_omp_clauses (gfc_code *code, gfc_omp_clauses *omp_clauses,
|
||||
break;
|
||||
}
|
||||
}
|
||||
if (code && code->op == EXEC_OMP_INTEROP)
|
||||
for (list = OMP_LIST_INIT; list <= OMP_LIST_DESTROY; list++)
|
||||
for (n = omp_clauses->lists[list]; n; n = n->next)
|
||||
{
|
||||
if (n->sym->ts.type != BT_INTEGER
|
||||
|| n->sym->ts.kind != gfc_index_integer_kind
|
||||
|| n->sym->attr.dimension
|
||||
|| n->sym->attr.flavor != FL_VARIABLE)
|
||||
gfc_error ("%qs at %L in %qs clause must be a scalar integer "
|
||||
"variable of %<omp_interop_kind%> kind", n->sym->name,
|
||||
&n->where, clause_names[list]);
|
||||
if (list != OMP_LIST_USE && n->sym->attr.intent == INTENT_IN)
|
||||
gfc_error ("%qs at %L in %qs clause must be definable",
|
||||
n->sym->name, &n->where, clause_names[list]);
|
||||
}
|
||||
|
||||
/* Detect specifically the case where we have "map(x) private(x)" and raise
|
||||
an error. If we have "...simd" combined directives though, the "private"
|
||||
|
||||
@ -2775,12 +2775,56 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses,
|
||||
case OMP_LIST_SCAN_EX:
|
||||
clause_code = OMP_CLAUSE_EXCLUSIVE;
|
||||
goto add_clause;
|
||||
case OMP_LIST_USE:
|
||||
clause_code = OMP_CLAUSE_USE;
|
||||
goto add_clause;
|
||||
case OMP_LIST_DESTROY:
|
||||
clause_code = OMP_CLAUSE_DESTROY;
|
||||
goto add_clause;
|
||||
|
||||
add_clause:
|
||||
omp_clauses
|
||||
= gfc_trans_omp_variable_list (clause_code, n, omp_clauses,
|
||||
declare_simd);
|
||||
break;
|
||||
|
||||
case OMP_LIST_INIT:
|
||||
{
|
||||
tree pref_type = NULL_TREE;
|
||||
const char *last = NULL;
|
||||
for (; n != NULL; n = n->next)
|
||||
if (n->sym->attr.referenced)
|
||||
{
|
||||
tree t = gfc_trans_omp_variable (n->sym, false);
|
||||
if (t == error_mark_node)
|
||||
continue;
|
||||
tree node = build_omp_clause (input_location,
|
||||
OMP_CLAUSE_INIT);
|
||||
OMP_CLAUSE_DECL (node) = t;
|
||||
if (n->u.init.target)
|
||||
OMP_CLAUSE_INIT_TARGET (node) = 1;
|
||||
if (n->u.init.targetsync)
|
||||
OMP_CLAUSE_INIT_TARGETSYNC (node) = 1;
|
||||
if (last != n->u2.init_interop)
|
||||
{
|
||||
last = n->u2.init_interop;
|
||||
if (n->u2.init_interop == NULL)
|
||||
pref_type = NULL_TREE;
|
||||
else
|
||||
{
|
||||
pref_type = build_string (n->u.init.len,
|
||||
n->u2.init_interop);
|
||||
TREE_TYPE (pref_type)
|
||||
= build_array_type_nelts (unsigned_char_type_node,
|
||||
n->u.init.len);
|
||||
}
|
||||
}
|
||||
OMP_CLAUSE_INIT_PREFER_TYPE (node) = pref_type;
|
||||
omp_clauses = gfc_trans_add_clause (node, omp_clauses);
|
||||
}
|
||||
break;
|
||||
}
|
||||
|
||||
case OMP_LIST_ALIGNED:
|
||||
for (; n != NULL; n = n->next)
|
||||
if (n->sym->attr.referenced || declare_simd)
|
||||
@ -8027,6 +8071,18 @@ gfc_trans_omp_target_update (gfc_code *code)
|
||||
return gfc_finish_block (&block);
|
||||
}
|
||||
|
||||
static tree
|
||||
gfc_trans_openmp_interop (gfc_code *code, gfc_omp_clauses *clauses)
|
||||
{
|
||||
stmtblock_t block;
|
||||
gfc_start_block (&block);
|
||||
tree omp_clauses = gfc_trans_omp_clauses (&block, clauses, code->loc);
|
||||
tree stmt = build1_loc (input_location, OMP_INTEROP, void_type_node,
|
||||
omp_clauses);
|
||||
gfc_add_expr_to_block (&block, stmt);
|
||||
return gfc_finish_block (&block);
|
||||
}
|
||||
|
||||
static tree
|
||||
gfc_trans_omp_workshare (gfc_code *code, gfc_omp_clauses *clauses)
|
||||
{
|
||||
@ -8365,8 +8421,7 @@ gfc_trans_omp_directive (gfc_code *code)
|
||||
case EXEC_OMP_WORKSHARE:
|
||||
return gfc_trans_omp_workshare (code, code->ext.omp_clauses);
|
||||
case EXEC_OMP_INTEROP:
|
||||
sorry ("%<!$OMP INTEROP%>");
|
||||
return build_empty_stmt (input_location);
|
||||
return gfc_trans_openmp_interop (code, code->ext.omp_clauses);
|
||||
default:
|
||||
gcc_unreachable ();
|
||||
}
|
||||
|
||||
@ -19300,6 +19300,11 @@ gimplify_expr (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p,
|
||||
break;
|
||||
}
|
||||
|
||||
case OMP_INTEROP:
|
||||
sorry_at (EXPR_LOCATION (*expr_p),
|
||||
"%<#pragma omp interop%> not yet supported");
|
||||
ret = GS_ERROR;
|
||||
break;
|
||||
case OMP_ATOMIC:
|
||||
case OMP_ATOMIC_READ:
|
||||
case OMP_ATOMIC_CAPTURE_OLD:
|
||||
|
||||
@ -29,7 +29,7 @@ along with GCC; see the file COPYING3. If not see
|
||||
extern bool omp_runtime_api_procname (const char *name);
|
||||
extern bool omp_runtime_api_call (const_tree fndecl);
|
||||
|
||||
extern int omp_get_fr_id_from_name (const char *);
|
||||
extern char omp_get_fr_id_from_name (const char *);
|
||||
extern const char *omp_get_name_from_fr_id (int);
|
||||
|
||||
#endif
|
||||
|
||||
@ -3501,7 +3501,7 @@ static const char* omp_interop_fr_str[] = {"cuda", "cuda_driver", "opencl",
|
||||
|
||||
/* Returns the foreign-runtime ID if found or 0 otherwise. */
|
||||
|
||||
int
|
||||
char
|
||||
omp_get_fr_id_from_name (const char *str)
|
||||
{
|
||||
static_assert (GOMP_INTEROP_IFR_LAST == ARRAY_SIZE (omp_interop_fr_str), "");
|
||||
@ -3509,7 +3509,7 @@ omp_get_fr_id_from_name (const char *str)
|
||||
for (unsigned i = 0; i < ARRAY_SIZE (omp_interop_fr_str); ++i)
|
||||
if (!strcmp (str, omp_interop_fr_str[i]))
|
||||
return i + 1;
|
||||
return 0;
|
||||
return GOMP_INTEROP_IFR_UNKNOWN;
|
||||
}
|
||||
|
||||
/* Returns the string value to a foreign-runtime integer value or NULL if value
|
||||
@ -3519,7 +3519,7 @@ const char *
|
||||
omp_get_name_from_fr_id (int fr_id)
|
||||
{
|
||||
if (fr_id < 1 || fr_id > (int) ARRAY_SIZE (omp_interop_fr_str))
|
||||
return NULL;
|
||||
return "<unknown>";
|
||||
return omp_interop_fr_str[fr_id-1];
|
||||
}
|
||||
|
||||
|
||||
119
gcc/testsuite/c-c++-common/gomp/interop-1.c
Normal file
119
gcc/testsuite/c-c++-common/gomp/interop-1.c
Normal file
@ -0,0 +1,119 @@
|
||||
/* { dg-do compile { target { c || c++11 } } } */
|
||||
/* { dg-additional-options "-std=c23" { target c } } */
|
||||
/* C++11 and C23 because of 'constexpr'. */
|
||||
|
||||
/* { dg-prune-output "sorry, unimplemented: '#pragma omp interop' not yet supported" } */
|
||||
|
||||
/* The following definitions are in omp_lib, which cannot be included
|
||||
in gcc/testsuite/ */
|
||||
|
||||
#if __cplusplus >= 201103L
|
||||
# define __GOMP_UINTPTR_T_ENUM : __UINTPTR_TYPE__
|
||||
#else
|
||||
# define __GOMP_UINTPTR_T_ENUM
|
||||
#endif
|
||||
|
||||
typedef enum omp_interop_t __GOMP_UINTPTR_T_ENUM
|
||||
{
|
||||
omp_interop_none = 0,
|
||||
__omp_interop_t_max__ = __UINTPTR_MAX__
|
||||
} omp_interop_t;
|
||||
|
||||
typedef enum omp_interop_fr_t
|
||||
{
|
||||
omp_ifr_cuda = 1,
|
||||
omp_ifr_cuda_driver = 2,
|
||||
omp_ifr_opencl = 3,
|
||||
omp_ifr_sycl = 4,
|
||||
omp_ifr_hip = 5,
|
||||
omp_ifr_level_zero = 6,
|
||||
omp_ifr_hsa = 7,
|
||||
omp_ifr_last = omp_ifr_hsa
|
||||
} omp_interop_fr_t;
|
||||
|
||||
|
||||
// ---------------------------------
|
||||
|
||||
void f()
|
||||
{
|
||||
constexpr omp_interop_fr_t ifr_scalar = omp_ifr_hsa;
|
||||
constexpr omp_interop_fr_t ifr_array[] = {omp_ifr_cuda, omp_ifr_hip};
|
||||
constexpr char my_string[] = "cuda";
|
||||
omp_interop_t obj1, obj2, obj3, obj4, obj5;
|
||||
int x;
|
||||
|
||||
#pragma omp interop init(obj1) init(target,targetsync : obj2, obj3) nowait // OK
|
||||
#pragma omp interop init(obj1) init (targetsync : obj2, obj3) nowait // OK
|
||||
#pragma omp interop init(obj1) init (targetsync , target : obj2, obj3) nowait // OK
|
||||
|
||||
#pragma omp interop init(obj1) init(target,targetsync,target: obj2, obj3) nowait // { dg-error "duplicate 'target' modifier" }
|
||||
#pragma omp interop init(obj1) init(target,targetsync, targetsync : obj2, obj3) nowait // { dg-error "duplicate 'targetsync' modifier" }
|
||||
|
||||
#pragma omp interop init(prefer_type("cuda", omp_ifr_opencl, omp_ifr_level_zero, "hsa"), targetsync : obj1) \
|
||||
destroy(obj2, obj3) depend(inout: x) use(obj4, obj5) device(device_num: 0)
|
||||
|
||||
#pragma omp interop init(prefer_type("cu" "da"), targetsync : obj1) // OK
|
||||
|
||||
#pragma omp assume contains(interop)
|
||||
{
|
||||
#pragma omp interop init(prefer_type("cuða") : obj3) // { dg-warning "unknown foreign runtime identifier 'cu\[^'\]*a'" }
|
||||
}
|
||||
|
||||
#pragma omp interop init(prefer_type("cu\0da") : obj3) // { dg-error "string literal must not contain '\\\\0'" }
|
||||
|
||||
#pragma omp interop depend(inout: x) , use(obj2), destroy(obj3) // OK, use or destroy might have 'targetsync'
|
||||
|
||||
#pragma omp interop depend(inout: x) use(obj2), destroy(obj3) // Likewise
|
||||
|
||||
#pragma omp interop depend(inout: x) init(targetsync : obj5) use(obj2), destroy(obj3) init(prefer_type("cuda"), targetsync : obj4) // OK
|
||||
|
||||
#pragma omp interop init ( target , prefer_type( { fr("hsa") }, "hip") : obj1) // { dg-error "expected '\{' before string constant" }
|
||||
|
||||
#pragma omp interop init ( target , prefer_type( { fr("hsa"), attr("ompx_nothing") , fr("hsa" ) }) :obj1) // { dg-error "duplicated 'fr' preference selector before '\\(' token" }
|
||||
|
||||
#pragma omp interop init ( prefer_type( 4, omp_ifr_hip*4) : obj1) // { dg-warning "unknown foreign runtime identifier '20'" }
|
||||
#pragma omp interop init ( prefer_type( __builtin_sin(3.3) : obj1)
|
||||
// { dg-error "'prefer_type' undeclared \\(first use in this function\\)" "" { target c } .-1 }
|
||||
// { dg-error "'prefer_type' has not been declared" "" { target c++ } .-2 }
|
||||
// { dg-error "expected '\\)' before '\\(' token" "" { target *-*-* } .-3 }
|
||||
|
||||
#pragma omp interop init ( prefer_type( __builtin_sin(3.3) ) : obj1) // { dg-error "expected string literal or constant integer expression before '\\)' token" }
|
||||
#pragma omp interop init ( prefer_type( {fr(4 ) }) : obj1) // OK
|
||||
#pragma omp interop init ( prefer_type( {fr("cu\0da" ) }) : obj1) // { dg-error "string literal must not contain '\\\\0'" }
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda\0") }) : obj1) // { dg-error "string literal must not contain '\\\\0'" }
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda" ) }) : obj1) // OK
|
||||
#pragma omp interop init ( prefer_type( {fr(omp_ifr_level_zero ) }, {fr(omp_ifr_hip)}) : obj1) // OK
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda", "cuda_driver") }) : obj1) // { dg-error "53: expected '\\)' before ',' token" }
|
||||
#pragma omp interop init ( prefer_type( {fr(my_string) }) : obj1) // { dg-error "56: expected string literal or constant integer expression before '\\)' token" }
|
||||
#pragma omp interop init ( prefer_type( {fr("hello" }) : obj1) // { dg-error "expected '\\)' before '\\(' token" }
|
||||
// { dg-error "'prefer_type' has not been declared" "" { target c++ } .-1 }
|
||||
#pragma omp interop init ( prefer_type( {fr("hello") }) : obj1)
|
||||
/* { dg-warning "unknown foreign runtime identifier 'hello' \\\[-Wopenmp\\\]" "" { target *-*-* } .-1 } */
|
||||
|
||||
#pragma omp interop init ( prefer_type( {fr(x) }) : obj1) // { dg-error "expected string literal or constant integer expression before '\\)' token" }
|
||||
|
||||
#pragma omp interop init ( prefer_type( {fr(ifr_scalar ) }) : obj1) // OK
|
||||
#pragma omp interop init ( prefer_type( {fr(ifr_array ) }) : obj1) // { dg-error "expected string literal or constant integer expression before '\\)' token" }
|
||||
// OK in C++, for C: constexpr arrays are not part of C23; however, they are/were under consideration for C2y.
|
||||
#pragma omp interop init ( prefer_type( {fr(ifr_array[0] ) }) : obj1)
|
||||
// { dg-error "expected string literal or constant integer expression before '\\)' token" "" { target c } .-1 }
|
||||
|
||||
#pragma omp interop init ( prefer_type( omp_ifr_level_zero, omp_ifr_hip ) : obj1) // OK
|
||||
#pragma omp interop init ( prefer_type( omp_ifr_level_zero +1 ) : obj1) // OK
|
||||
#pragma omp interop init ( prefer_type( x ) : obj1) // { dg-error "expected string literal or constant integer expression before '\\)' token" }
|
||||
|
||||
#pragma omp interop init ( prefer_type( ifr_scalar ) : obj1) // OK
|
||||
#pragma omp interop init ( prefer_type( ifr_array ) : obj1) // { dg-error "expected string literal or constant integer expression before '\\)' token" }
|
||||
// OK in C++, for C: constexpr arrays are not part of C23; however, they are/were under consideration for C2y.
|
||||
#pragma omp interop init ( prefer_type( ifr_array[1] ) : obj1)
|
||||
// { dg-error "expected string literal or constant integer expression before '\\)' token" "" { target c } .-1 }
|
||||
|
||||
#pragma omp interop init ( prefer_type( 4, omp_ifr_hip*4) : obj1) // { dg-warning "unknown foreign runtime identifier '20'" }
|
||||
#pragma omp interop init ( prefer_type( 4, 1, 3) : obj1)
|
||||
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda") }, {fr(omp_ifr_hsa)} , {attr("ompx_a") } , {fr(omp_ifr_hip) }) : obj1)
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda") }, {fr(omp_ifr_hsa,omp_ifr_level_zero)} , {attr("ompx_a") } , {fr(omp_ifr_hip) }) : obj1) // { dg-error "73: expected '\\)' before ',' token" }
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda",5) }, {fr(omp_ifr_hsa,omp_ifr_level_zero)} , {attr("ompx_a") } , {fr(omp_ifr_hip) }) : obj1) // { dg-error "53: expected '\\)' before ',' token" }
|
||||
#pragma omp interop init ( prefer_type( {fr("sycl"), attr("ompx_1", "ompx_2"), attr("ompx_3") }, {attr("ompx_4", "ompx_5"),fr(omp_ifr_level_zero)} ) : obj1)
|
||||
#pragma omp interop init ( prefer_type( { fr(5), attr("ompx_1") }, {fr(omp_ifr_hsa)} , {attr("ompx_a") } ) : obj1)
|
||||
}
|
||||
127
gcc/testsuite/c-c++-common/gomp/interop-2.c
Normal file
127
gcc/testsuite/c-c++-common/gomp/interop-2.c
Normal file
@ -0,0 +1,127 @@
|
||||
/* { dg-do compile { target { c || c++11 } } } */
|
||||
/* { dg-additional-options "-std=c23" { target c } } */
|
||||
/* C++11 and C23 because of 'constexpr'. */
|
||||
|
||||
/* { dg-prune-output "sorry, unimplemented: '#pragma omp interop' not yet supported" } */
|
||||
|
||||
/* The following definitions are in omp_lib, which cannot be included
|
||||
in gcc/testsuite/ */
|
||||
|
||||
#if __cplusplus >= 201103L
|
||||
# define __GOMP_UINTPTR_T_ENUM : __UINTPTR_TYPE__
|
||||
#else
|
||||
# define __GOMP_UINTPTR_T_ENUM
|
||||
#endif
|
||||
|
||||
typedef enum omp_interop_t __GOMP_UINTPTR_T_ENUM
|
||||
{
|
||||
omp_interop_none = 0,
|
||||
__omp_interop_t_max__ = __UINTPTR_MAX__
|
||||
} omp_interop_t;
|
||||
|
||||
typedef enum omp_interop_fr_t
|
||||
{
|
||||
omp_ifr_cuda = 1,
|
||||
omp_ifr_cuda_driver = 2,
|
||||
omp_ifr_opencl = 3,
|
||||
omp_ifr_sycl = 4,
|
||||
omp_ifr_hip = 5,
|
||||
omp_ifr_level_zero = 6,
|
||||
omp_ifr_hsa = 7,
|
||||
omp_ifr_last = omp_ifr_hsa
|
||||
} omp_interop_fr_t;
|
||||
|
||||
|
||||
// ---------------------------------
|
||||
|
||||
void f(const omp_interop_t ocp)
|
||||
{
|
||||
constexpr omp_interop_t oce = omp_interop_none;
|
||||
const omp_interop_t occ = omp_interop_none;
|
||||
omp_interop_t od[5];
|
||||
omp_interop_t *op;
|
||||
short o2;
|
||||
float of;
|
||||
|
||||
#pragma omp interop init (ocp) // { dg-error "'ocp' shall not be const" }
|
||||
#pragma omp interop init (oce) // { dg-error "'oce' shall not be const" }
|
||||
#pragma omp interop init (occ) // { dg-error "'occ' shall not be const" }
|
||||
#pragma omp interop init (od) // { dg-error "'od' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop init (od[1])// { dg-error "expected '\\)' before '\\\[' token" }
|
||||
// { dg-error "'od' must be of 'omp_interop_t'" "" { target *-*-* } .-1 }
|
||||
#pragma omp interop init (op) // { dg-error "'op' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop init (*op)
|
||||
// { dg-error "expected identifier before '\\*' token" "" { target c } .-1 }
|
||||
// { dg-error "expected unqualified-id before '\\*' token" "" { target c++ } .-2 }
|
||||
#pragma omp interop init (o2) // { dg-error "'o2' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop init (of) // { dg-error "'of' must be of 'omp_interop_t'" }
|
||||
|
||||
#pragma omp interop use (ocp) // OK
|
||||
#pragma omp interop use (oce) // odd but okay
|
||||
#pragma omp interop use (occ) // okayish
|
||||
#pragma omp interop use (od) // { dg-error "'od' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop use (od[1])// { dg-error "expected '\\)' before '\\\[' token" }
|
||||
// { dg-error "'od' must be of 'omp_interop_t'" "" { target *-*-* } .-1 }
|
||||
#pragma omp interop use (op) // { dg-error "'op' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop use (*op)
|
||||
// { dg-error "expected identifier before '\\*' token" "" { target c } .-1 }
|
||||
// { dg-error "expected unqualified-id before '\\*' token" "" { target c++ } .-2 }
|
||||
#pragma omp interop use (o2) // { dg-error "'o2' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop use (of) // { dg-error "'of' must be of 'omp_interop_t'" }
|
||||
|
||||
#pragma omp interop destroy (ocp) // { dg-error "'ocp' shall not be const" }
|
||||
#pragma omp interop destroy (oce) // { dg-error "'oce' shall not be const" }
|
||||
#pragma omp interop destroy (occ) // { dg-error "'occ' shall not be const" }
|
||||
#pragma omp interop destroy (od) // { dg-error "'od' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop destroy (od[1])// { dg-error "expected '\\)' before '\\\[' token" }
|
||||
// { dg-error "'od' must be of 'omp_interop_t'" "" { target *-*-* } .-1 }
|
||||
#pragma omp interop destroy (op) // { dg-error "'op' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop destroy (*op)
|
||||
// { dg-error "expected identifier before '\\*' token" "" { target c } .-1 }
|
||||
// { dg-error "expected unqualified-id before '\\*' token" "" { target c++ } .-2 }
|
||||
#pragma omp interop destroy (o2) // { dg-error "'o2' must be of 'omp_interop_t'" }
|
||||
#pragma omp interop destroy (of) // { dg-error "'of' must be of 'omp_interop_t'" }
|
||||
}
|
||||
|
||||
void g()
|
||||
{
|
||||
omp_interop_t obj1, obj2, obj3, obj4, obj5;
|
||||
int x;
|
||||
|
||||
#pragma omp interop init ( prefer_type( {fr("") }) : obj1) // { dg-error "non-empty string literal expected before '\\)' token" }
|
||||
#pragma omp interop init ( prefer_type( {fr("hip") , attr(omp_ifr_cuda) }) : obj1) ! { dg-error "expected string literal before 'omp_ifr_cuda'" }
|
||||
|
||||
#pragma omp interop init ( prefer_type( {fr("hip") , attr("myooption") }) : obj1) // { dg-error "'attr' string literal must start with 'ompx_'" }
|
||||
#pragma omp interop init ( prefer_type( {fr("hip") , attr("ompx_option") , attr("ompx_") } ) : obj1)
|
||||
#pragma omp interop init ( prefer_type( {fr("hip") , attr("ompx_option") }, { attr("ompx_") } ) : obj1)
|
||||
#pragma omp interop init ( prefer_type( {fr("hip") , attr("ompx_option") } { attr("ompx_") } ) : obj1) // { dg-error "expected '\\)' or ',' before '\{' token" }
|
||||
#pragma omp interop init ( prefer_type( {fr("hip") , attr("ompx_option") ) : obj1)
|
||||
// { dg-error "expected ',' or '\}' before '\\)' token" "" { target c } .-1 }
|
||||
// { dg-error "prefer_type' has not been declared" "" { target c++ } .-2 }
|
||||
// { dg-error "expected '\\)' before '\\(' token" "" { target c++ } .-3 }
|
||||
|
||||
#pragma omp interop init ( prefer_type( {fr("hip") attr("ompx_option") ) : obj1)
|
||||
// { dg-error "expected ',' or '\}' before 'attr'" "" { target c } .-1 }
|
||||
// { dg-error "prefer_type' has not been declared" "" { target c++ } .-2 }
|
||||
// { dg-error "expected '\\)' before '\\(' token" "" { target c++ } .-3 }
|
||||
#pragma omp interop init ( prefer_type( {fr("hip")}), prefer_type("cuda") : obj1) // { dg-error "duplicate 'prefer_type' modifier" }
|
||||
|
||||
#pragma omp interop init ( prefer_type( {attr("ompx_option1,ompx_option2") } ) : obj1) // { dg-error "'attr' string literal must not contain a comma" }
|
||||
|
||||
#pragma omp interop init ( prefer_type( {attr("ompx_option1,ompx_option2") ) : obj1)
|
||||
// { dg-error "'attr' string literal must not contain a comma" "" { target c } .-1 }
|
||||
// { dg-error "prefer_type' has not been declared" "" { target c++ } .-2 }
|
||||
// { dg-error "expected '\\)' before '\\(' token" "" { target c++ } .-3 }
|
||||
|
||||
#pragma omp interop init ( targetsync other ) : obj1)
|
||||
// { dg-error "'targetsync' undeclared \\(first use in this function\\)" "" { target c } .-1 }
|
||||
// { dg-error "'targetsync' has not been declared" "" { target c++ } .-2 }
|
||||
// { dg-error "expected '\\)' before 'other'" "" { target *-*-* } .-3 }
|
||||
// { dg-error "expected an OpenMP clause before ':' token" "" { target *-*-* } .-4 }
|
||||
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda") } ), other : obj1) // { dg-error "'init' clause with modifier other than 'prefer_type', 'target' or 'targetsync' before 'other'" }
|
||||
#pragma omp interop init ( prefer_type( {fr("cuda") } ), obj1)
|
||||
// { dg-error "'prefer_type' undeclared \\(first use in this function\\)" "" { target c } .-1 }
|
||||
// { dg-error "'prefer_type' has not been declared" "" { target c++ } .-2 }
|
||||
// { dg-error "expected '\\)' before '\\(' token" "" { target *-*-* } .-3 }
|
||||
}
|
||||
82
gcc/testsuite/c-c++-common/gomp/interop-3.c
Normal file
82
gcc/testsuite/c-c++-common/gomp/interop-3.c
Normal file
@ -0,0 +1,82 @@
|
||||
/* { dg-prune-output "sorry, unimplemented: '#pragma omp interop' not yet supported" } */
|
||||
|
||||
/* The following definitions are in omp_lib, which cannot be included
|
||||
in gcc/testsuite/ */
|
||||
|
||||
#if __cplusplus >= 201103L
|
||||
# define __GOMP_UINTPTR_T_ENUM : __UINTPTR_TYPE__
|
||||
#else
|
||||
# define __GOMP_UINTPTR_T_ENUM
|
||||
#endif
|
||||
|
||||
typedef enum omp_interop_t __GOMP_UINTPTR_T_ENUM
|
||||
{
|
||||
omp_interop_none = 0,
|
||||
__omp_interop_t_max__ = __UINTPTR_MAX__
|
||||
} omp_interop_t;
|
||||
|
||||
typedef enum omp_interop_fr_t
|
||||
{
|
||||
omp_ifr_cuda = 1,
|
||||
omp_ifr_cuda_driver = 2,
|
||||
omp_ifr_opencl = 3,
|
||||
omp_ifr_sycl = 4,
|
||||
omp_ifr_hip = 5,
|
||||
omp_ifr_level_zero = 6,
|
||||
omp_ifr_hsa = 7,
|
||||
omp_ifr_last = omp_ifr_hsa
|
||||
} omp_interop_fr_t;
|
||||
|
||||
|
||||
// ---------------------------------
|
||||
|
||||
void f()
|
||||
{
|
||||
omp_interop_t obj1, obj2, obj3, obj4, obj5;
|
||||
omp_interop_t target, targetsync, prefer_type;
|
||||
int x;
|
||||
|
||||
#pragma omp interop init(obj1) init(target,targetsync : obj2, obj3) nowait
|
||||
|
||||
#pragma omp interop init(prefer_type("cuda", omp_ifr_opencl, omp_ifr_level_zero, "hsa"), targetsync : obj1) \
|
||||
destroy(obj2, obj3) depend(inout: x) use(obj4, obj5) device(device_num: 0)
|
||||
|
||||
#pragma omp assume contains(interop)
|
||||
{
|
||||
#pragma omp interop init(prefer_type("cu da") : obj3) // { dg-warning "unknown foreign runtime identifier 'cu da'" }
|
||||
}
|
||||
|
||||
#pragma omp interop init(obj1, obj2, obj1), use(obj4) destroy(obj4)
|
||||
// { dg-error "'obj4' appears more than once in action clauses" "" { target *-*-* } .-1 }
|
||||
// { dg-error "'obj1' appears more than once in action clauses" "" { target *-*-* } .-2 }
|
||||
|
||||
#pragma omp interop depend(inout: x) // { dg-error "'depend' clause requires action clauses with 'targetsync' interop-type" }
|
||||
|
||||
#pragma omp interop depend(inout: x) , use(obj2), destroy(obj3) // OK, use or destroy might have 'targetsync'
|
||||
|
||||
#pragma omp interop depend(inout: x) use(obj2), destroy(obj3) // Likewise
|
||||
|
||||
#pragma omp interop depend(inout: x) use(obj2), destroy(obj3) init(obj4) // { dg-error "'depend' clause requires action clauses with 'targetsync' interop-type" }
|
||||
// { dg-note "69: 'init' clause lacks the 'targetsync' modifier" "" { target c } .-1 }
|
||||
// { dg-note "70: 'init' clause lacks the 'targetsync' modifier" "" { target c++ } .-2 }
|
||||
|
||||
#pragma omp interop depend(inout: x) init(targetsync : obj5) use(obj2), destroy(obj3) init(obj4) // { dg-error "'depend' clause requires action clauses with 'targetsync' interop-type" }
|
||||
// { dg-note "'init' clause lacks the 'targetsync' modifier" "" { target *-*-* } .-1 }
|
||||
#pragma omp interop depend(inout: x) init(targetsync : obj5) use(obj2), destroy(obj3) init(prefer_type("cuda"), targetsync : obj4) // OK
|
||||
|
||||
#pragma omp interop init(target, targetsync, prefer_type, obj1)
|
||||
#pragma omp interop init(prefer_type, obj1, target, targetsync)
|
||||
|
||||
// Duplicated variable name or duplicated modifier:
|
||||
#pragma omp interop init(target, targetsync,target : obj1) // { dg-error "duplicate 'target' modifier" }
|
||||
#pragma omp interop init(target, targetsync,target) // { dg-error "'target' appears more than once in action clauses" }
|
||||
#pragma omp interop init(target : target, targetsync,target) // { dg-error "'target' appears more than once in action clauses" }
|
||||
|
||||
#pragma omp interop init(target, targetsync,targetsync : obj1) // { dg-error "duplicate 'targetsync' modifier" }
|
||||
#pragma omp interop init(target, targetsync,targetsync) // { dg-error "targetsync' appears more than once in action clause" }
|
||||
#pragma omp interop init(target : target, targetsync,targetsync) // { dg-error "targetsync' appears more than once in action clause" }
|
||||
|
||||
#pragma omp interop init(, targetsync, prefer_type, obj1, target)
|
||||
// { dg-error "expected identifier before ',' token" "" { target c } .-1 }
|
||||
// { dg-error "expected unqualified-id before ',' token" "" { target c++ } .-2 }
|
||||
}
|
||||
75
gcc/testsuite/c-c++-common/gomp/interop-4.c
Normal file
75
gcc/testsuite/c-c++-common/gomp/interop-4.c
Normal file
@ -0,0 +1,75 @@
|
||||
/* { dg-additional-options "-fdump-tree-original" } */
|
||||
|
||||
/* The following definitions are in omp_lib, which cannot be included
|
||||
in gcc/testsuite/ */
|
||||
|
||||
#if __cplusplus >= 201103L
|
||||
# define __GOMP_UINTPTR_T_ENUM : __UINTPTR_TYPE__
|
||||
#else
|
||||
# define __GOMP_UINTPTR_T_ENUM
|
||||
#endif
|
||||
|
||||
typedef enum omp_interop_t __GOMP_UINTPTR_T_ENUM
|
||||
{
|
||||
omp_interop_none = 0,
|
||||
__omp_interop_t_max__ = __UINTPTR_MAX__
|
||||
} omp_interop_t;
|
||||
|
||||
typedef enum omp_interop_fr_t
|
||||
{
|
||||
omp_ifr_cuda = 1,
|
||||
omp_ifr_cuda_driver = 2,
|
||||
omp_ifr_opencl = 3,
|
||||
omp_ifr_sycl = 4,
|
||||
omp_ifr_hip = 5,
|
||||
omp_ifr_level_zero = 6,
|
||||
omp_ifr_hsa = 7,
|
||||
omp_ifr_last = omp_ifr_hsa
|
||||
} omp_interop_fr_t;
|
||||
|
||||
void
|
||||
f()
|
||||
{
|
||||
omp_interop_t obj1, obj2, obj3, obj4, obj5, obj6, obj7;
|
||||
int x[6];
|
||||
|
||||
#pragma omp interop init ( obj1, obj2) use (obj3) destroy(obj4) init(obj5) destroy(obj6) use(obj7) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/* { dg-final { scan-tree-dump-times "#pragma omp interop use\\(obj7\\) destroy\\(obj6\\) init\\(obj5\\) destroy\\(obj4\\) use\\(obj3\\) init\\(obj2\\) init\\(obj1\\)\[\r\n\]" 1 "original" } } */
|
||||
|
||||
#pragma omp interop nowait init (targetsync : obj1, obj2) use (obj3) destroy(obj4) init(target, targetsync : obj5) destroy(obj6) use(obj7) depend(inout: x) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/* { dg-final { scan-tree-dump-times "#pragma omp interop depend\\(inout:x\\) use\\(obj7\\) destroy\\(obj6\\) init\\(target, targetsync: obj5\\) destroy\\(obj4\\) use\\(obj3\\) init\\(targetsync: obj2\\) init\\(targetsync: obj1\\) nowait\[\r\n\]" 1 "original" } } */
|
||||
|
||||
#pragma omp interop init ( obj1, obj2) init (target: obj3) init(targetsync : obj4) init(target,targetsync: obj5) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/* { dg-final { scan-tree-dump-times "#pragma omp interop init\\(target, targetsync: obj5\\) init\\(targetsync: obj4\\) init\\(target: obj3\\) init\\(obj2\\) init\\(obj1\\)\[\r\n\]" 1 "original" } } */
|
||||
|
||||
/* -------------------------------------------- */
|
||||
|
||||
#pragma omp interop init (target, prefer_type(omp_ifr_cuda, omp_ifr_cuda+1, "hsa", "myPrivateInterop", omp_ifr_cuda-2) : obj1, obj2) init (target: obj3) init(prefer_type(omp_ifr_hip, "sycl", omp_ifr_opencl), targetsync : obj4, obj7) init(target,prefer_type("level_zero", omp_ifr_level_zero+0),targetsync: obj5) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/*
|
||||
{ dg-warning "unknown foreign runtime identifier 'myPrivateInterop' \\\[-Wopenmp\\\]" "" { target *-*-* } .-2 }
|
||||
{ dg-warning "unknown foreign runtime identifier '-1' \\\[-Wopenmp\\\]" "" { target *-*-* } .-3 }
|
||||
|
||||
{ dg!final { scan-tree-dump-times "#pragma omp interop init\\(prefer_type\\({fr\\(\"level_zero\"\\)}, {fr\\(\"level_zero\"\\)}\\), target, targetsync: obj5\\) init\\(prefer_type\\({fr\\(\"hip\"\\)}, {fr\\(\"sycl\"\\)}, {fr\\(\"opencl\"\\)}\\), targetsync: obj7\\) init\\(prefer_type\\({fr\\(\"hip\"\\)}, {fr\\(\"sycl\"\\)}, {fr\\(\"opencl\"\\)}\\), targetsync: obj4\\) init\\(target: obj3\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\)}, {fr\\(\"hsa\"\\)}, {fr\\(\"<unknown>\"\\)}, {fr\\(\"<unknown>\"\\)}\\), target: obj2\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\)}, {fr\\(\"hsa\"\\)}, {fr\\(\"<unknown>\"\\)}, {fr\\(\"<unknown>\"\\)}\\), target: obj1\\)\[\r\n\]" 1 "original" } }
|
||||
*/
|
||||
|
||||
|
||||
/* -------------------------------------------- */
|
||||
|
||||
#pragma omp interop init ( target, prefer_type( {fr("hip"), attr("ompx_gnu_prio:1", "ompx_gnu_debug")}, {attr("ompx_gnu_nicest"), attr("ompx_something")}) : obj1, obj2) init ( prefer_type( {fr("cuda")}, {fr(omp_ifr_cuda_driver), attr("ompx_nix")}, {fr("best")}), targetsync : obj3, obj4) nowait use(obj5) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/*
|
||||
{ dg-warning "unknown foreign runtime identifier 'best' \\\[-Wopenmp\\\]" "" { target *-*-* } .-2 }
|
||||
|
||||
{ dg-final { scan-tree-dump-times "#pragma omp interop use\\(obj5\\) nowait init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\),attr\\(\"ompx_nix\"\\)}, {fr\\(\"<unknown>\"\\)}\\), targetsync: obj4\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\),attr\\(\"ompx_nix\"\\)}, {fr\\(\"<unknown>\"\\)}\\), targetsync: obj3\\) init\\(prefer_type\\({fr\\(\"hip\"\\),attr\\(\"ompx_gnu_prio:1\"\\),attr\\(\"ompx_gnu_debug\"\\)}, {attr\\(\"ompx_gnu_nicest\"\\),attr\\(\"ompx_something\"\\)}\\), target: obj2\\) init\\(prefer_type\\({fr\\(\"hip\"\\),attr\\(\"ompx_gnu_prio:1\"\\),attr\\(\"ompx_gnu_debug\"\\)}, {attr\\(\"ompx_gnu_nicest\"\\),attr\\(\"ompx_something\"\\)}\\), target: obj1\\)\[\r\n\]" 1 "original" } }
|
||||
*/
|
||||
|
||||
}
|
||||
|
||||
void
|
||||
g (int *y)
|
||||
{
|
||||
omp_interop_t io1, io2, io3, io4, io5;
|
||||
|
||||
[[omp::directive (interop,init(prefer_type({fr("level_zero")}, {fr(omp_ifr_sycl),attr("ompx_in_order"),attr("ompx_queue:in_order")}), targetsync : io1, io2),use(io3),destroy(io4,io5),depend(inout:y),nowait)]]; /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
|
||||
/* { dg-final { scan-tree-dump-times "#pragma omp interop nowait depend\\(inout:y\\) destroy\\(io5\\) destroy\\(io4\\) use\\(io3\\) init\\(prefer_type\\(\{fr\\(\"level_zero\"\\)\}, \{fr\\(\"sycl\"\\),attr\\(\"ompx_in_order\"\\),attr\\(\"ompx_queue:in_order\"\\)\}\\), targetsync: io2\\) init\\(prefer_type\\(\{fr\\(\"level_zero\"\\)\}, \{fr\\(\"sycl\"\\),attr\\(\"ompx_in_order\"\\),attr\\(\"ompx_queue:in_order\"\\)\}\\), targetsync: io1\\)\[\r\n\]" 1 "original" } } */
|
||||
}
|
||||
90
gcc/testsuite/g++.dg/gomp/interop-5.C
Normal file
90
gcc/testsuite/g++.dg/gomp/interop-5.C
Normal file
@ -0,0 +1,90 @@
|
||||
/* { dg-do compile { target c++11 } } */
|
||||
/* { dg-additional-options "-fdump-tree-original" } */
|
||||
|
||||
/* { dg-prune-output "sorry, unimplemented: '#pragma omp interop' not yet supported" } */
|
||||
|
||||
/* The following definitions are in omp_lib, which cannot be included
|
||||
in gcc/testsuite/ */
|
||||
|
||||
#if __cplusplus >= 201103L
|
||||
# define __GOMP_UINTPTR_T_ENUM : __UINTPTR_TYPE__
|
||||
#else
|
||||
# define __GOMP_UINTPTR_T_ENUM
|
||||
#endif
|
||||
|
||||
typedef enum omp_interop_t __GOMP_UINTPTR_T_ENUM
|
||||
{
|
||||
omp_interop_none = 0,
|
||||
__omp_interop_t_max__ = __UINTPTR_MAX__
|
||||
} omp_interop_t;
|
||||
|
||||
typedef enum omp_interop_fr_t
|
||||
{
|
||||
omp_ifr_cuda = 1,
|
||||
omp_ifr_cuda_driver = 2,
|
||||
omp_ifr_opencl = 3,
|
||||
omp_ifr_sycl = 4,
|
||||
omp_ifr_hip = 5,
|
||||
omp_ifr_level_zero = 6,
|
||||
omp_ifr_hsa = 7,
|
||||
omp_ifr_last = omp_ifr_hsa
|
||||
} omp_interop_fr_t;
|
||||
|
||||
template<typename T, typename T2, typename T3>
|
||||
void
|
||||
f ()
|
||||
{
|
||||
T obj1, obj2, obj3, obj4, obj5, obj6, obj7;
|
||||
T2 x[6];
|
||||
constexpr T3 ifr_hip = omp_ifr_hip;
|
||||
constexpr T3 ifr_cuda = omp_ifr_cuda;
|
||||
constexpr T3 ifr_cuda_driver = omp_ifr_cuda_driver;
|
||||
constexpr T3 ifr_opencl = omp_ifr_opencl;
|
||||
constexpr T3 ifr_level_zero = (T3) (omp_ifr_sycl + 2);
|
||||
constexpr T3 ifr_invalid = (T3) 99;
|
||||
|
||||
#pragma omp interop init ( obj1, obj2) use (obj3) destroy(obj4) init(obj5) destroy(obj6) use(obj7) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/* { dg-final { scan-tree-dump-times "#pragma omp interop use\\(obj7\\) destroy\\(obj6\\) init\\(obj5\\) destroy\\(obj4\\) use\\(obj3\\) init\\(obj2\\) init\\(obj1\\)\[\r\n\]" 2 "original" } } */
|
||||
|
||||
#pragma omp interop nowait init (targetsync : obj1, obj2) use (obj3) destroy(obj4) init(target, targetsync : obj5) destroy(obj6) use(obj7) depend(inout: x) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/* { dg-final { scan-tree-dump-times "#pragma omp interop depend\\(inout:x\\) use\\(obj7\\) destroy\\(obj6\\) init\\(target, targetsync: obj5\\) destroy\\(obj4\\) use\\(obj3\\) init\\(targetsync: obj2\\) init\\(targetsync: obj1\\) nowait\[\r\n\]" 2 "original" } } */
|
||||
|
||||
#pragma omp interop init ( obj1, obj2) init (target: obj3) init(targetsync : obj4) init(target,targetsync: obj5) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/* { dg-final { scan-tree-dump-times "#pragma omp interop init\\(target, targetsync: obj5\\) init\\(targetsync: obj4\\) init\\(target: obj3\\) init\\(obj2\\) init\\(obj1\\)\[\r\n\]" 2 "original" } } */
|
||||
|
||||
/* -------------------------------------------- */
|
||||
|
||||
#pragma omp interop init (target, prefer_type(ifr_invalid, 123, ifr_cuda, ifr_cuda+1, "hsa", "myPrivateInterop", ifr_cuda-2) : obj1, obj2)
|
||||
/*
|
||||
{ dg-warning "49: unknown foreign runtime identifier '99' \\\[-Wopenmp\\\]" "" { target *-*-* } .-2 }
|
||||
{ dg-warning "62: unknown foreign runtime identifier '123' \\\[-Wopenmp\\\]" "" { target *-*-* } .-3 }
|
||||
{ dg-warning "96: unknown foreign runtime identifier 'myPrivateInterop' \\\[-Wopenmp\\\]" "" { target *-*-* } .-4 }
|
||||
{ dg-warning "124: unknown foreign runtime identifier '-1' \\\[-Wopenmp\\\]" "" { target *-*-* } .-5 }
|
||||
|
||||
{ dg-final { scan-tree-dump-times "#pragma omp interop init\\(prefer_type\\(\{fr\\(\"<unknown>\"\\)\}, \{fr\\(\"<unknown>\"\\)\}, \{fr\\(\"cuda\"\\)\}, \{fr\\(\"cuda_driver\"\\)\}, \{fr\\(\"hsa\"\\)\}, \{fr\\(\"<unknown>\"\\)\}, \{fr\\(\"<unknown>\"\\)\}\\), target: obj2\\) init\\(prefer_type\\(\{fr\\(\"<unknown>\"\\)\}, \{fr\\(\"<unknown>\"\\)\}, \{fr\\(\"cuda\"\\)\}, \{fr\\(\"cuda_driver\"\\)\}, \{fr\\(\"hsa\"\\)\}, \{fr\\(\"<unknown>\"\\)\}, \{fr\\(\"<unknown>\"\\)\}\\), target: obj1\\)\[\r\n\]" 2 "original" } }
|
||||
*/
|
||||
|
||||
#pragma omp interop init (target, prefer_type(ifr_cuda, ifr_cuda+1, "hsa", "myPrivateInterop", ifr_cuda-2) : obj1, obj2) init (target: obj3) init(prefer_type(ifr_hip, "sycl", ifr_opencl), targetsync : obj4, obj7) init(target,prefer_type("level_zero", ifr_level_zero+0),targetsync: obj5) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/*
|
||||
{ dg-warning "unknown foreign runtime identifier 'myPrivateInterop' \\\[-Wopenmp\\\]" "" { target *-*-* } .-2 }
|
||||
{ dg-warning "unknown foreign runtime identifier '-1' \\\[-Wopenmp\\\]" "" { target *-*-* } .-3 }
|
||||
|
||||
{ dg-final { scan-tree-dump-times "#pragma omp interop init\\(prefer_type\\({fr\\(\"level_zero\"\\)}, {fr\\(\"level_zero\"\\)}\\), target, targetsync: obj5\\) init\\(prefer_type\\({fr\\(\"hip\"\\)}, {fr\\(\"sycl\"\\)}, {fr\\(\"opencl\"\\)}\\), targetsync: obj7\\) init\\(prefer_type\\({fr\\(\"hip\"\\)}, {fr\\(\"sycl\"\\)}, {fr\\(\"opencl\"\\)}\\), targetsync: obj4\\) init\\(target: obj3\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\)}, {fr\\(\"hsa\"\\)}, {fr\\(\"<unknown>\"\\)}, {fr\\(\"<unknown>\"\\)}\\), target: obj2\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\)}, {fr\\(\"hsa\"\\)}, {fr\\(\"<unknown>\"\\)}, {fr\\(\"<unknown>\"\\)}\\), target: obj1\\)\[\r\n\]" 2 "original" } }
|
||||
*/
|
||||
|
||||
/* -------------------------------------------- */
|
||||
|
||||
#pragma omp interop init ( target, prefer_type( {fr("hip"), attr("ompx_gnu_prio:1", "ompx_gnu_debug")}, {attr("ompx_gnu_nicest"), attr("ompx_something")}) : obj1, obj2) init ( prefer_type( {fr("cuda")}, {fr(ifr_cuda_driver), attr("ompx_nix")}, {fr("best")}), targetsync : obj3, obj4) nowait use(obj5) /* { dg-message "'#pragma omp interop' not yet supported" } */
|
||||
/*
|
||||
{ dg-warning "unknown foreign runtime identifier 'best' \\\[-Wopenmp\\\]" "" { target *-*-* } .-2 }
|
||||
|
||||
{ dg-final { scan-tree-dump-times "#pragma omp interop use\\(obj5\\) nowait init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\),attr\\(\"ompx_nix\"\\)}, {fr\\(\"<unknown>\"\\)}\\), targetsync: obj4\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\),attr\\(\"ompx_nix\"\\)}, {fr\\(\"<unknown>\"\\)}\\), targetsync: obj3\\) init\\(prefer_type\\({fr\\(\"hip\"\\),attr\\(\"ompx_gnu_prio:1\"\\),attr\\(\"ompx_gnu_debug\"\\)}, {attr\\(\"ompx_gnu_nicest\"\\),attr\\(\"ompx_something\"\\)}\\), target: obj2\\) init\\(prefer_type\\({fr\\(\"hip\"\\),attr\\(\"ompx_gnu_prio:1\"\\),attr\\(\"ompx_gnu_debug\"\\)}, {attr\\(\"ompx_gnu_nicest\"\\),attr\\(\"ompx_something\"\\)}\\), target: obj1\\)\[\r\n\]" 2 "original" } }
|
||||
*/
|
||||
}
|
||||
|
||||
void
|
||||
g (int *y)
|
||||
{
|
||||
f<omp_interop_t, int, omp_interop_fr_t> ();
|
||||
f<omp_interop_t, int, int> ();
|
||||
}
|
||||
@ -33,18 +33,27 @@ integer(omp_interop_fr_kind), parameter :: ifr_array(2) = [omp_ifr_cuda, omp_ifr
|
||||
integer(omp_interop_kind) :: obj1, obj2, obj3, obj4, obj5
|
||||
integer :: x
|
||||
|
||||
!$omp interop init(obj1) init(target,targetsync,target,targetsync : obj2, obj3) nowait
|
||||
!$omp interop init(obj1) init(target,targetsync : obj2, obj3) nowait ! OK
|
||||
!$omp interop init(obj1) init (targetsync : obj2, obj3) nowait ! OK
|
||||
!$omp interop init(obj1) init (targetsync , target : obj2, obj3) nowait ! OK
|
||||
|
||||
!$omp interop init(prefer_type("cu"//"da", omp_ifr_opencl, omp_ifr_level_zero, "hsa"), targetsync : obj1) &
|
||||
!$omp interop init(obj1) init(target,targetsync,target: obj2, obj3) nowait ! { dg-error "Duplicate 'target'" }
|
||||
!$omp interop init(obj1) init(target,targetsync, targetsync : obj2, obj3) nowait ! { dg-error "Duplicate 'targetsync'" }
|
||||
|
||||
!$omp interop init(prefer_type("cuda", omp_ifr_opencl, omp_ifr_level_zero, "hsa"), targetsync : obj1) &
|
||||
!$omp& destroy(obj2, obj3) depend(inout: x) use(obj4, obj5) device(device_num: 0)
|
||||
|
||||
!$omp interop init(prefer_type("cu" // "da"), targetsync : obj1) ! { dg-error "37: Expected ',' or '\\)'" }
|
||||
! { dg-warning "Unknown foreign runtime identifier 'cu' at \\(1\\) \\\[-Wopenmp\\\]" "" { target *-*-* } .-1 }
|
||||
|
||||
!$omp assume contains(interop)
|
||||
!$omp interop init(prefer_type("cu"//char(1)//"da") : obj3) ! { dg-warning "Unknown foreign runtime identifier 'cu\\\\x01da'" }
|
||||
!$omp interop init(prefer_type("cuða") : obj3) ! { dg-warning "Unknown foreign runtime identifier 'cu\[^'\]*a'" }
|
||||
!$omp end assume
|
||||
|
||||
!$omp interop init(prefer_type("cu"//char(0)//"da") : obj3) ! { dg-error "Unexpected null character in character literal" }
|
||||
!$omp interop init(prefer_type("cu"//char(0)//"da") : obj3) ! { dg-error "36: Expected ',' or '\\)'" }
|
||||
! { dg-warning "Unknown foreign runtime identifier 'cu' at \\(1\\) \\\[-Wopenmp\\\]" "" { target *-*-* } .-1 }
|
||||
|
||||
!$omp interop depend(inout: x) , use(obj2), destroy(obj3) ! OK, use or destory might have 'targetsync'
|
||||
!$omp interop depend(inout: x) , use(obj2), destroy(obj3) ! OK, use or destroy might have 'targetsync'
|
||||
|
||||
!$omp interop depend(inout: x) use(obj2), destroy(obj3) ! Likewise
|
||||
|
||||
@ -56,15 +65,19 @@ integer :: x
|
||||
|
||||
!$omp interop init ( prefer_type( 4, omp_ifr_hip*4) : obj1) ! { dg-warning "Unknown foreign runtime identifier '20'" }
|
||||
!$omp interop init ( prefer_type( sin(3.3) : obj1) ! { dg-error "Expected constant scalar integer expression or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(4 ) }) : obj1) ! { dg-error "Expected scalar integer parameter or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(4_"cuda" ) }) : obj1) ! { dg-error "Expected scalar integer parameter or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(4 ) }) : obj1) ! OK
|
||||
!$omp interop init ( prefer_type( {fr(4_"cuda" ) }) : obj1) ! { dg-error "Expected constant scalar integer expression or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(c_char_"cuda") }) : obj1) ! OK
|
||||
!$omp interop init ( prefer_type( {fr(1_"cuda" ) }) : obj1) ! OK
|
||||
!$omp interop init ( prefer_type( {fr(omp_ifr_level_zero ) }, {fr(omp_ifr_hip)}) : obj1) ! OK
|
||||
!$omp interop init ( prefer_type( {fr(omp_ifr_level_zero + 1) }) : obj1) ! { dg-error "Expected scalar integer parameter or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(x) }) : obj1) ! { dg-error "Expected scalar integer parameter or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(ifr_array ) }) : obj1) ! { dg-error "Expected scalar integer parameter or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(ifr_array(1) ) }) : obj1) ! { dg-error "Expected scalar integer parameter or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr("cuda" // "_driver") }) : obj1) ! { dg-error "46: Expected '\\)'" }
|
||||
!$omp interop init ( prefer_type( {fr(trim("cuda" // "_driver")) }) : obj1) ! { dg-error "38: Expected constant scalar integer expression or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr("hello" }) : obj1) ! { dg-error "47: Expected '\\)'" }
|
||||
! { dg-warning "Unknown foreign runtime identifier 'hello' at \\(1\\) \\\[-Wopenmp\\\]" "" { target *-*-* } .-1 }
|
||||
|
||||
!$omp interop init ( prefer_type( {fr(x) }) : obj1) ! { dg-error "Expected constant scalar integer expression or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(ifr_array ) }) : obj1) ! { dg-error "Expected constant scalar integer expression or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(ifr_array(1) ) }) : obj1)
|
||||
|
||||
!$omp interop init ( prefer_type( omp_ifr_level_zero, omp_ifr_hip ) : obj1) ! OK
|
||||
!$omp interop init ( prefer_type( omp_ifr_level_zero +1 ) : obj1) ! OK
|
||||
@ -75,8 +88,10 @@ integer :: x
|
||||
!$omp interop init ( prefer_type( 4, omp_ifr_hip*4) : obj1) ! { dg-warning "Unknown foreign runtime identifier '20'" }
|
||||
!$omp interop init ( prefer_type( 4, 1, 3) : obj1)
|
||||
|
||||
!$omp interop init ( prefer_type( {fr("cuda","sycl") }, {fr(omp_ifr_hsa,omp_ifr_level_zero)} , {attr("ompx_a") } , {fr(omp_ifr_hip) }) : obj1)
|
||||
!$omp interop init ( prefer_type( {fr("cuda","sycl"), attr("ompx_1", "ompx_2"), attr("ompx_3") }, {attr("ompx_4", "ompx_5"),fr(omp_ifr_hsa,omp_ifr_level_zero)} ) : obj1)
|
||||
!$omp interop init ( prefer_type( { fr("cuda","sycl"), attr("ompx_1") }, {fr(omp_ifr_hsa,omp_ifr_level_zero)} , {attr("ompx_a") } ) : obj1)
|
||||
!$omp interop init ( prefer_type( {fr("cuda") }, {fr(omp_ifr_hsa)} , {attr("ompx_a") } , {fr(omp_ifr_hip) }) : obj1)
|
||||
!$omp interop init ( prefer_type( {fr("cuda") }, {fr(omp_ifr_hsa,omp_ifr_level_zero)} , {attr("ompx_a") } , {fr(omp_ifr_hip) }) : obj1) ! { dg-error "65: Expected '\\)'" }
|
||||
!$omp interop init ( prefer_type( {fr("cuda",5) }, {fr(omp_ifr_hsa,omp_ifr_level_zero)} , {attr("ompx_a") } , {fr(omp_ifr_hip) }) : obj1) ! { dg-error "45: Expected '\\)' at" }
|
||||
!$omp interop init ( prefer_type( {fr("sycl"), attr("ompx_1", "ompx_2"), attr("ompx_3") }, {attr("ompx_4", "ompx_5"),fr(omp_ifr_level_zero)} ) : obj1)
|
||||
!$omp interop init ( prefer_type( { fr(5), attr("ompx_1") }, {fr(omp_ifr_hsa)} , {attr("ompx_a") } ) : obj1)
|
||||
|
||||
end
|
||||
|
||||
@ -17,16 +17,50 @@ module m
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_hsa = 7
|
||||
end module m
|
||||
|
||||
subroutine s(ointent)
|
||||
use m
|
||||
implicit none
|
||||
integer(omp_interop_kind), parameter :: op = 0
|
||||
integer(omp_interop_kind),intent(in) :: ointent
|
||||
integer(omp_interop_kind) :: od(5)
|
||||
integer(1) :: o1
|
||||
integer, parameter :: mykind = mod (omp_interop_kind, 100) ! remove saving the 'comes from c_int' info
|
||||
real(mykind) :: or
|
||||
|
||||
!$omp interop init (op) ! { dg-error "'op' at \\(1\\) in 'INIT' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
! { dg-error "Object 'op' is not a variable at \\(1\\)" "" { target *-*-* } .-1 }
|
||||
!$omp interop init (ointent) ! { dg-error "'ointent' at \\(1\\) in 'INIT' clause must be definable" }
|
||||
!$omp interop init (od) ! { dg-error "'od' at \\(1\\) in 'INIT' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
!$omp interop init (od(1)) ! { dg-error "Syntax error in OpenMP variable list" }
|
||||
!$omp interop init (o1) ! { dg-error "'o1' at \\(1\\) in 'INIT' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
!$omp interop init (or) ! { dg-error "'or' at \\(1\\) in 'INIT' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
|
||||
!$omp interop use (op) ! { dg-error "'op' at \\(1\\) in 'USE' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
! { dg-error "Object 'op' is not a variable at \\(1\\)" "" { target *-*-* } .-1 }
|
||||
!$omp interop use (ointent) ! okay
|
||||
!$omp interop use (od) ! { dg-error "'od' at \\(1\\) in 'USE' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
!$omp interop use (od(1)) ! { dg-error "Syntax error in OpenMP variable list" }
|
||||
!$omp interop use (o1) ! { dg-error "'o1' at \\(1\\) in 'USE' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
!$omp interop use (or) ! { dg-error "'or' at \\(1\\) in 'USE' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
|
||||
!$omp interop destroy (op) ! { dg-error "'op' at \\(1\\) in 'DESTROY' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
! { dg-error "Object 'op' is not a variable at \\(1\\)" "" { target *-*-* } .-1 }
|
||||
!$omp interop destroy (ointent) ! { dg-error "'ointent' at \\(1\\) in 'DESTROY' clause must be definable" }
|
||||
!$omp interop destroy (od) ! { dg-error "'od' at \\(1\\) in 'DESTROY' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
!$omp interop destroy (od(1)) ! { dg-error "Syntax error in OpenMP variable list" }
|
||||
!$omp interop destroy (o1) ! { dg-error "'o1' at \\(1\\) in 'DESTROY' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
!$omp interop destroy (or) ! { dg-error "'or' at \\(1\\) in 'DESTROY' clause must be a scalar integer variable of 'omp_interop_kind' kind" }
|
||||
|
||||
end subroutine
|
||||
|
||||
program main
|
||||
use m
|
||||
implicit none
|
||||
|
||||
!$omp requires reverse_offload
|
||||
|
||||
integer(omp_interop_kind) :: obj1, obj2, obj3, obj4, obj5
|
||||
integer :: x
|
||||
|
||||
!$omp interop init ( prefer_type( {fr(1_"") }) : obj1) ! { dg-error "Expected scalar integer parameter or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(1_"") }) : obj1) ! { dg-error "Expected constant scalar integer expression or non-empty default-kind character literal" }
|
||||
!$omp interop init ( prefer_type( {fr(1_"hip") , attr(omp_ifr_cuda) }) : obj1) ! { dg-error "Expected default-kind character literal" }
|
||||
|
||||
!$omp interop init ( prefer_type( {fr(1_"hip") , attr("myooption") }) : obj1) ! { dg-error "Character literal at .1. must start with 'ompx_'" }
|
||||
|
||||
@ -21,19 +21,17 @@ program main
|
||||
use m
|
||||
implicit none
|
||||
|
||||
!$omp requires reverse_offload
|
||||
|
||||
integer(omp_interop_kind) :: obj1, obj2, obj3, obj4, obj5
|
||||
integer(omp_interop_kind) :: target, targetsync,prefer_type
|
||||
integer :: x
|
||||
|
||||
!$omp interop init(obj1) init(target,targetsync,target,targetsync : obj2, obj3) nowait
|
||||
!$omp interop init(obj1) init(target,targetsync : obj2, obj3) nowait
|
||||
|
||||
!$omp interop init(prefer_type("cu"//"da", omp_ifr_opencl, omp_ifr_level_zero, "hsa"), targetsync : obj1) &
|
||||
!$omp interop init(prefer_type(1_"cuda", omp_ifr_opencl, omp_ifr_level_zero, "hsa"), targetsync : obj1) &
|
||||
!$omp& destroy(obj2, obj3) depend(inout: x) use(obj4, obj5) device(device_num: 0)
|
||||
|
||||
!$omp assume contains(interop)
|
||||
!$omp interop init(prefer_type("cu"//char(1)//"da") : obj3) ! { dg-warning "Unknown foreign runtime identifier 'cu\\\\x01da'" }
|
||||
!$omp interop init(prefer_type("cu da") : obj3) ! { dg-warning "Unknown foreign runtime identifier 'cu da'" }
|
||||
!$omp end assume
|
||||
|
||||
!$omp interop init(obj1, obj2, obj1), use(obj4) destroy(obj4)
|
||||
@ -42,7 +40,7 @@ integer :: x
|
||||
|
||||
!$omp interop depend(inout: x) ! { dg-error "DEPEND clause at .1. requires action clause with 'targetsync' interop-type" }
|
||||
|
||||
!$omp interop depend(inout: x) , use(obj2), destroy(obj3) ! OK, use or destory might have 'targetsync'
|
||||
!$omp interop depend(inout: x) , use(obj2), destroy(obj3) ! OK, use or destroy might have 'targetsync'
|
||||
|
||||
!$omp interop depend(inout: x) use(obj2), destroy(obj3) ! Likewise
|
||||
|
||||
@ -53,7 +51,16 @@ integer :: x
|
||||
|
||||
!$omp interop init(target, targetsync, prefer_type, obj1)
|
||||
!$omp interop init(prefer_type, obj1, target, targetsync)
|
||||
!$omp interop init(target, targetsync,target) ! { dg-error "Symbol 'target' present on multiple clauses" }
|
||||
|
||||
! Duplicated variable name or duplicated modifier:
|
||||
!$omp interop init(target, targetsync,target : obj1) ! { dg-error "Duplicate 'target' at \\(1\\)" }
|
||||
!$omp interop init(target, targetsync,target) ! { dg-error "Duplicate 'target' at \\(1\\)" }
|
||||
!$omp interop init(target : target, targetsync,target) ! { dg-error "Symbol 'target' present on multiple clauses" }
|
||||
|
||||
!$omp interop init(target, targetsync,targetsync : obj1) ! { dg-error "Duplicate 'targetsync' at \\(1\\)" }
|
||||
!$omp interop init(target, targetsync,targetsync) ! { dg-error "Duplicate 'targetsync' at \\(1\\)" }
|
||||
!$omp interop init(target : target, targetsync,targetsync) ! { dg-error "Symbol 'targetsync' present on multiple clauses" }
|
||||
|
||||
|
||||
!$omp interop init(, targetsync, prefer_type, obj1, target) ! { dg-error "Syntax error in OpenMP variable list" }
|
||||
end
|
||||
|
||||
56
gcc/testsuite/gfortran.dg/gomp/interop-4.f90
Normal file
56
gcc/testsuite/gfortran.dg/gomp/interop-4.f90
Normal file
@ -0,0 +1,56 @@
|
||||
! { dg-additional-options "-fdump-tree-original" }
|
||||
|
||||
module m
|
||||
use iso_c_binding
|
||||
implicit none
|
||||
|
||||
! The following definitions are in omp_lib, which cannot be included
|
||||
! in gcc/testsuite/
|
||||
integer, parameter :: omp_interop_kind = c_intptr_t
|
||||
integer, parameter :: omp_interop_fr_kind = c_int
|
||||
|
||||
integer (omp_interop_kind), parameter :: omp_interop_none = 0_omp_interop_kind
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_cuda = 1
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_cuda_driver = 2
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_opencl = 3
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_sycl = 4
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_hip = 5
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_level_zero = 6
|
||||
integer (omp_interop_fr_kind), parameter :: omp_ifr_hsa = 7
|
||||
end module m
|
||||
|
||||
subroutine s
|
||||
use m
|
||||
implicit none
|
||||
|
||||
integer(omp_interop_kind) :: obj1, obj2, obj3, obj4, obj5, obj6, obj7
|
||||
integer :: x(6)
|
||||
|
||||
!$omp interop init ( obj1, obj2) use (obj3) destroy(obj4) init(obj5) destroy(obj6) use(obj7) ! { dg-message "'#pragma omp interop' not yet supported" }
|
||||
! { dg-final { scan-tree-dump-times "#pragma omp interop init\\(obj1\\) init\\(obj2\\) init\\(obj5\\) use\\(obj3\\) use\\(obj7\\) destroy\\(obj4\\) destroy\\(obj6\\)\[\r\n\]" 1 "original" } }
|
||||
|
||||
!$omp interop nowait init (targetsync : obj1, obj2) use (obj3) destroy(obj4) init(target, targetsync : obj5) destroy(obj6) use(obj7) depend(inout: x) ! { dg-message "'#pragma omp interop' not yet supported" }
|
||||
! { dg-final { scan-tree-dump-times "#pragma omp interop depend\\(inout:x\\) init\\(targetsync: obj1\\) init\\(targetsync: obj2\\) init\\(target, targetsync: obj5\\) use\\(obj3\\) use\\(obj7\\) destroy\\(obj4\\) destroy\\(obj6\\) nowait\[\r\n\]" 1 "original" } }
|
||||
|
||||
!$omp interop init ( obj1, obj2) init (target: obj3) init(targetsync : obj4) init(target,targetsync: obj5) ! { dg-message "'#pragma omp interop' not yet supported" }
|
||||
! { dg-final { scan-tree-dump-times "#pragma omp interop init\\(obj1\\) init\\(obj2\\) init\\(target: obj3\\) init\\(targetsync: obj4\\) init\\(target, targetsync: obj5\\)\[\r\n\]" 1 "original" } }
|
||||
|
||||
! --------------------------------------------
|
||||
|
||||
!$omp interop init (target, prefer_type(omp_ifr_cuda, omp_ifr_cuda+1, "hsa", "myPrivateInterop", omp_ifr_cuda-2) : obj1, obj2) init (target: obj3) init(prefer_type(omp_ifr_hip, "sycl", omp_ifr_opencl), targetsync : obj4, obj7) init(target,prefer_type("level_zero", omp_ifr_level_zero+0),targetsync: obj5) ! { dg-message "'#pragma omp interop' not yet supported" }
|
||||
!
|
||||
! { dg-warning "Unknown foreign runtime identifier 'myPrivateInterop' at \\(1\\) \\\[-Wopenmp\\\]" "" { target *-*-* } .-2 }
|
||||
! { dg-warning "Unknown foreign runtime identifier '-1' at \\(1\\) \\\[-Wopenmp\\\]" "" { target *-*-* } .-3 }
|
||||
!
|
||||
! { dg-final { scan-tree-dump-times "#pragma omp interop init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\)}, {fr\\(\"hsa\"\\)}, {fr\\(\"<unknown>\"\\)}, {fr\\(\"<unknown>\"\\)}\\), target: obj1\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\)}, {fr\\(\"hsa\"\\)}, {fr\\(\"<unknown>\"\\)}, {fr\\(\"<unknown>\"\\)}\\), target: obj2\\) init\\(target: obj3\\) init\\(prefer_type\\({fr\\(\"hip\"\\)}, {fr\\(\"sycl\"\\)}, {fr\\(\"opencl\"\\)}\\), targetsync: obj4\\) init\\(prefer_type\\({fr\\(\"hip\"\\)}, {fr\\(\"sycl\"\\)}, {fr\\(\"opencl\"\\)}\\), targetsync: obj7\\) init\\(prefer_type\\({fr\\(\"level_zero\"\\)}, {fr\\(\"level_zero\"\\)}\\), target, targetsync: obj5\\)\[\r\n\]" 1 "original" } }
|
||||
|
||||
|
||||
! --------------------------------------------
|
||||
|
||||
!$omp interop init ( target, prefer_type( {fr(1_"hip"), attr("ompx_gnu_prio:1", 1_"ompx_gnu_debug")}, {attr("ompx_gnu_nicest"), attr("ompx_something")}) : obj1, obj2) init ( prefer_type( {fr("cuda")}, {fr(omp_ifr_cuda_driver), attr("ompx_nix")}, {fr("best")}), targetsync : obj3, obj4) nowait use(obj5) ! { dg-message "'#pragma omp interop' not yet supported" }
|
||||
!
|
||||
! ! { dg-warning "Unknown foreign runtime identifier 'best' at \\(1\\) \\\[-Wopenmp\\\]" "" { target *-*-* } .-2 }
|
||||
!
|
||||
! { dg-final { scan-tree-dump-times "#pragma omp interop init\\(prefer_type\\({fr\\(\"hip\"\\),attr\\(\"ompx_gnu_prio:1\"\\),attr\\(\"ompx_gnu_debug\"\\)}, {attr\\(\"ompx_gnu_nicest\"\\),attr\\(\"ompx_something\"\\)}\\), target: obj1\\) init\\(prefer_type\\({fr\\(\"hip\"\\),attr\\(\"ompx_gnu_prio:1\"\\),attr\\(\"ompx_gnu_debug\"\\)}, {attr\\(\"ompx_gnu_nicest\"\\),attr\\(\"ompx_something\"\\)}\\), target: obj2\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\),attr\\(\"ompx_nix\"\\)}, {fr\\(\"<unknown>\"\\)}\\), targetsync: obj3\\) init\\(prefer_type\\({fr\\(\"cuda\"\\)}, {fr\\(\"cuda_driver\"\\),attr\\(\"ompx_nix\"\\)}, {fr\\(\"<unknown>\"\\)}\\), targetsync: obj4\\) use\\(obj5\\) nowait\[\r\n\]" 1 "original" } }
|
||||
|
||||
end
|
||||
@ -374,6 +374,19 @@ enum omp_clause_code {
|
||||
|
||||
/* Range END above for: OMP_CLAUSE_SIZE */
|
||||
|
||||
/* OpenMP clause: destroy (variable-list ). */
|
||||
OMP_CLAUSE_DESTROY,
|
||||
|
||||
/* Range START below for: OMP_CLAUSE_INIT_PREFER_TYPE */
|
||||
|
||||
/* OpenMP clause: init ( [modifier-list : ] variable-list ). */
|
||||
OMP_CLAUSE_INIT,
|
||||
|
||||
/* Range END above for: OMP_CLAUSE_INIT_PREFER_TYPE */
|
||||
|
||||
/* OpenMP clause: use (variable-list ). */
|
||||
OMP_CLAUSE_USE,
|
||||
|
||||
/* OpenACC clause: gang [(gang-argument-list)].
|
||||
Where
|
||||
gang-argument-list: [gang-argument-list, ] gang-argument
|
||||
|
||||
@ -452,6 +452,49 @@ dump_omp_iterators (pretty_printer *pp, tree iter, int spc, dump_flags_t flags)
|
||||
pp_right_paren (pp);
|
||||
}
|
||||
|
||||
/* Dump OpenMP's prefer_type of the init clause. */
|
||||
|
||||
static void
|
||||
dump_omp_init_prefer_type (pretty_printer *pp, tree t)
|
||||
{
|
||||
if (t == NULL_TREE)
|
||||
return;
|
||||
pp_string (pp, "prefer_type(");
|
||||
const char *str = TREE_STRING_POINTER (t);
|
||||
while (str[0] == (char) GOMP_INTEROP_IFR_SEPARATOR)
|
||||
{
|
||||
bool has_fr = false;
|
||||
pp_character (pp, '{');
|
||||
str++;
|
||||
while (str[0] != (char) GOMP_INTEROP_IFR_SEPARATOR)
|
||||
{
|
||||
if (has_fr)
|
||||
pp_character (pp, ',');
|
||||
has_fr = true;
|
||||
pp_string (pp, "fr(\"");
|
||||
pp_string (pp, omp_get_name_from_fr_id (str[0]));
|
||||
pp_string (pp, "\")");
|
||||
str++;
|
||||
}
|
||||
str++;
|
||||
if (has_fr && str[0] != '\0')
|
||||
pp_character (pp, ',');
|
||||
while (str[0] != '\0')
|
||||
{
|
||||
pp_string (pp, "attr(\"");
|
||||
pp_string (pp, str);
|
||||
pp_string (pp, "\")");
|
||||
str += strlen (str) + 1;
|
||||
if (str[0] != '\0')
|
||||
pp_character (pp, ',');
|
||||
}
|
||||
str++;
|
||||
pp_character (pp, '}');
|
||||
if (str[0] != '\0')
|
||||
pp_string (pp, ", ");
|
||||
}
|
||||
pp_right_paren (pp);
|
||||
}
|
||||
|
||||
/* Dump OMP clause CLAUSE, without following OMP_CLAUSE_CHAIN.
|
||||
|
||||
@ -605,6 +648,44 @@ dump_omp_clause (pretty_printer *pp, tree clause, int spc, dump_flags_t flags)
|
||||
pp_right_paren (pp);
|
||||
break;
|
||||
|
||||
case OMP_CLAUSE_DESTROY:
|
||||
pp_string (pp, "destroy(");
|
||||
dump_generic_node (pp, OMP_CLAUSE_DECL (clause),
|
||||
spc, flags, false);
|
||||
pp_right_paren (pp);
|
||||
break;
|
||||
|
||||
case OMP_CLAUSE_INIT:
|
||||
pp_string (pp, "init(");
|
||||
dump_omp_init_prefer_type (pp, OMP_CLAUSE_INIT_PREFER_TYPE (clause));
|
||||
if (OMP_CLAUSE_INIT_TARGET (clause))
|
||||
{
|
||||
if (OMP_CLAUSE_INIT_PREFER_TYPE (clause))
|
||||
pp_string (pp, ", ");
|
||||
pp_string (pp, "target");
|
||||
}
|
||||
if (OMP_CLAUSE_INIT_TARGETSYNC (clause))
|
||||
{
|
||||
if (OMP_CLAUSE_INIT_PREFER_TYPE (clause) || OMP_CLAUSE_INIT_TARGET (clause))
|
||||
pp_string (pp, ", ");
|
||||
pp_string (pp, "targetsync");
|
||||
}
|
||||
if (OMP_CLAUSE_INIT_PREFER_TYPE (clause)
|
||||
|| OMP_CLAUSE_INIT_TARGET (clause)
|
||||
|| OMP_CLAUSE_INIT_TARGETSYNC (clause))
|
||||
pp_string (pp, ": ");
|
||||
dump_generic_node (pp, OMP_CLAUSE_DECL (clause),
|
||||
spc, flags, false);
|
||||
pp_right_paren (pp);
|
||||
break;
|
||||
|
||||
case OMP_CLAUSE_USE:
|
||||
pp_string (pp, "use(");
|
||||
dump_generic_node (pp, OMP_CLAUSE_DECL (clause),
|
||||
spc, flags, false);
|
||||
pp_right_paren (pp);
|
||||
break;
|
||||
|
||||
case OMP_CLAUSE_SELF:
|
||||
pp_string (pp, "self(");
|
||||
dump_generic_node (pp, OMP_CLAUSE_SELF_EXPR (clause),
|
||||
@ -3991,6 +4072,12 @@ dump_generic_node (pretty_printer *pp, tree node, int spc, dump_flags_t flags,
|
||||
dump_omp_clauses (pp, OMP_DISPATCH_CLAUSES (node), spc, flags);
|
||||
goto dump_omp_body;
|
||||
|
||||
case OMP_INTEROP:
|
||||
pp_string (pp, "#pragma omp interop");
|
||||
dump_omp_clauses (pp, OMP_INTEROP_CLAUSES (node), spc, flags);
|
||||
is_expr = false;
|
||||
break;
|
||||
|
||||
case OMP_SECTION:
|
||||
pp_string (pp, "#pragma omp section");
|
||||
goto dump_omp_body;
|
||||
|
||||
@ -271,6 +271,9 @@ unsigned const char omp_clause_num_ops[] =
|
||||
1, /* OMP_CLAUSE_HAS_DEVICE_ADDR */
|
||||
1, /* OMP_CLAUSE_DOACROSS */
|
||||
2, /* OMP_CLAUSE__CACHE_ */
|
||||
1, /* OMP_CLAUSE_DESTROY */
|
||||
2, /* OMP_CLAUSE_INIT */
|
||||
1, /* OMP_CLAUSE_USE */
|
||||
2, /* OMP_CLAUSE_GANG */
|
||||
1, /* OMP_CLAUSE_ASYNC */
|
||||
1, /* OMP_CLAUSE_WAIT */
|
||||
@ -369,6 +372,9 @@ const char * const omp_clause_code_name[] =
|
||||
"has_device_addr",
|
||||
"doacross",
|
||||
"_cache_",
|
||||
"destroy",
|
||||
"init",
|
||||
"use",
|
||||
"gang",
|
||||
"async",
|
||||
"wait",
|
||||
|
||||
@ -1246,7 +1246,7 @@ DEFTREECODE (OMP_TILE, "omp_tile", tcc_statement, 7)
|
||||
Operands like for OMP_FOR. */
|
||||
DEFTREECODE (OMP_UNROLL, "omp_unroll", tcc_statement, 7)
|
||||
|
||||
/* OpenMP - #pragma acc loop [clause1 ... clauseN]
|
||||
/* OpenACC - #pragma acc loop [clause1 ... clauseN]
|
||||
Operands like for OMP_FOR. */
|
||||
DEFTREECODE (OACC_LOOP, "oacc_loop", tcc_statement, 7)
|
||||
|
||||
@ -1311,6 +1311,10 @@ DEFTREECODE (OMP_SCAN, "omp_scan", tcc_statement, 2)
|
||||
Operand 1: OMP_DISPATCH_CLAUSES: List of clauses. */
|
||||
DEFTREECODE (OMP_DISPATCH, "omp_dispatch", tcc_statement, 2)
|
||||
|
||||
/* OpenMP - #pragma omp interop [clause1 ... clauseN]
|
||||
Operand 0: OMP_INTEROP_CLAUSES: List of clauses. */
|
||||
DEFTREECODE (OMP_INTEROP, "omp_inteorp", tcc_statement, 1)
|
||||
|
||||
/* OpenMP - #pragma omp section
|
||||
Operand 0: OMP_SECTION_BODY: Section body. */
|
||||
DEFTREECODE (OMP_SECTION, "omp_section", tcc_statement, 1)
|
||||
|
||||
12
gcc/tree.h
12
gcc/tree.h
@ -1555,6 +1555,9 @@ class auto_suppress_location_wrappers
|
||||
#define OMP_FOR_PRE_BODY(NODE) TREE_OPERAND (OMP_LOOPING_CHECK (NODE), 5)
|
||||
#define OMP_FOR_ORIG_DECLS(NODE) TREE_OPERAND (OMP_LOOPING_CHECK (NODE), 6)
|
||||
|
||||
#define OMP_INTEROP_CLAUSES(NODE)\
|
||||
TREE_OPERAND (OMP_INTEROP_CHECK (NODE), 0)
|
||||
|
||||
#define OMP_LOOPXFORM_CHECK(NODE) TREE_RANGE_CHECK (NODE, OMP_TILE, OMP_UNROLL)
|
||||
#define OMP_LOOPXFORM_LOWERED(NODE) \
|
||||
(OMP_LOOPXFORM_CHECK (NODE)->base.public_flag)
|
||||
@ -1831,6 +1834,15 @@ class auto_suppress_location_wrappers
|
||||
#define OMP_CLAUSE_MOTION_PRESENT(NODE) \
|
||||
(OMP_CLAUSE_RANGE_CHECK (NODE, OMP_CLAUSE_FROM, OMP_CLAUSE_TO)->base.deprecated_flag)
|
||||
|
||||
#define OMP_CLAUSE_INIT_TARGET(NODE) \
|
||||
(OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_INIT)->base.public_flag)
|
||||
#define OMP_CLAUSE_INIT_TARGETSYNC(NODE) \
|
||||
(OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_INIT)->base.deprecated_flag)
|
||||
#define OMP_CLAUSE_INIT_PREFER_TYPE(NODE) \
|
||||
OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (OMP_CLAUSE_CHECK (NODE), \
|
||||
OMP_CLAUSE_INIT, \
|
||||
OMP_CLAUSE_INIT), 1)
|
||||
|
||||
/* Nonzero if this map clause is for array (rather than pointer) based array
|
||||
section with zero bias. Both the non-decl OMP_CLAUSE_MAP and corresponding
|
||||
OMP_CLAUSE_MAP with GOMP_MAP_POINTER are marked with this flag. */
|
||||
|
||||
@ -400,10 +400,11 @@ enum gomp_map_kind
|
||||
#define GOMP_REQUIRES_TARGET_USED 0x200
|
||||
#define GOMP_REQUIRES_SELF_MAPS 0x400
|
||||
|
||||
/* Interop foreign-runtime data. */
|
||||
/* Interop foreign-runtime data;
|
||||
OpenMP defines positive values; reserve 0 and negative for GCC. */
|
||||
#define GOMP_INTEROP_IFR_LAST 7
|
||||
#define GOMP_INTEROP_IFR_SEPARATOR -1
|
||||
#define GOMP_INTEROP_IFR_NONE -2
|
||||
#define GOMP_INTEROP_IFR_SEPARATOR ((char)(-__INT8_MAX__-1))
|
||||
#define GOMP_INTEROP_IFR_UNKNOWN ((char)(-__INT8_MAX__))
|
||||
|
||||
/* HSA specific data structures. */
|
||||
|
||||
|
||||
Loading…
Reference in New Issue
Block a user