diff options
115 files changed, 6083 insertions, 124 deletions
diff --git a/gcc/ChangeLog b/gcc/ChangeLog index d7e8702..59f447c 100644 --- a/gcc/ChangeLog +++ b/gcc/ChangeLog @@ -1,3 +1,26 @@ +2025-05-29 Yuta Mukai <mukai.yuta@fujitsu.com> + + Backported from master: + 2025-05-28 Yuta Mukai <mukai.yuta@fujitsu.com> + + * config/aarch64/aarch64-cores.def (fujitsu-monaka): Update ISA + features. + +2025-05-27 Eric Botcazou <ebotcazou@adacore.com> + + * ipa-sra.cc (scan_expr_access): Also disqualify storage order + barriers from splitting. + * tree.h (storage_order_barrier_p): Also return false if the + operand of the VIEW_CONVERT_EXPR has reverse storage order. + +2025-05-27 Jonathan Wakely <jwakely@redhat.com> + + Backported from master: + 2025-05-27 Jonathan Wakely <jwakely@redhat.com> + + * doc/extend.texi (Common Variable Attributes): Fix typo in + description of nonstring. + 2025-05-25 Michael J. Eager <eager@eagercon.com> PR target/86772 diff --git a/gcc/ChangeLog.omp b/gcc/ChangeLog.omp index eaeb97f..b832b2a 100644 --- a/gcc/ChangeLog.omp +++ b/gcc/ChangeLog.omp @@ -1,3 +1,12 @@ +2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + + Backported from master: + 2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + Richard Biener <rguenther@suse.de> + + PR middle-end/119835 + * tree-nrv.cc (pass_nrv::execute): Defuse 'RESULT_DECL' check. + 2025-05-22 Thomas Schwinge <tschwinge@baylibre.com> Backported from master: diff --git a/gcc/DATESTAMP b/gcc/DATESTAMP index dbf258b..5646e6e 100644 --- a/gcc/DATESTAMP +++ b/gcc/DATESTAMP @@ -1 +1 @@ -20250526 +20250602 diff --git a/gcc/c/ChangeLog b/gcc/c/ChangeLog index d842028..061f1c6 100644 --- a/gcc/c/ChangeLog +++ b/gcc/c/ChangeLog @@ -1,3 +1,29 @@ +2025-05-30 Sandra Loosemore <sloosemore@baylibre.com> + + Backported from master: + 2025-05-29 Sandra Loosemore <sloosemore@baylibre.com> + + * c-parser.cc (c_parser_skip_to_closing_brace): New, copied from + the equivalent function in the C++ front end. + (c_parser_skip_to_end_of_block_or_statement): Pass false to + the error flag. + (c_parser_omp_context_selector): Immediately return error_mark_node + after giving an error that the integer trait property is invalid, + similarly to C++ front end. + (c_parser_omp_context_selector_specification): Likewise handle + error return from c_parser_omp_context_selector similarly to C++. + (c_parser_omp_metadirective): Do not call + c_parser_skip_to_end_of_block_or_statement after an error. + +2025-05-30 Sandra Loosemore <sloosemore@baylibre.com> + + Backported from master: + 2025-05-29 Sandra Loosemore <sloosemore@baylibre.com> + + PR c/120180 + * c-parser.cc (c_parser_omp_metadirective): Only consume the + token if it is the expected close paren. + 2025-05-02 Jakub Jelinek <jakub@redhat.com> Backported from master: diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc index 9af7440..d132704 100644 --- a/gcc/c/c-parser.cc +++ b/gcc/c/c-parser.cc @@ -1418,6 +1418,51 @@ c_parser_skip_to_end_of_parameter (c_parser *parser) parser->error = false; } +/* Skip tokens until a non-nested closing curly brace is the next + token, or there are no more tokens. Return true in the first case, + false otherwise. */ + +static bool +c_parser_skip_to_closing_brace (c_parser *parser) +{ + unsigned nesting_depth = 0; + + while (true) + { + c_token *token = c_parser_peek_token (parser); + + switch (token->type) + { + case CPP_PRAGMA_EOL: + if (!parser->in_pragma) + break; + /* FALLTHRU */ + case CPP_EOF: + /* If we've run out of tokens, stop. */ + return false; + + case CPP_CLOSE_BRACE: + /* If the next token is a non-nested `}', then we have reached + the end of the current block. */ + if (nesting_depth-- == 0) + return true; + break; + + case CPP_OPEN_BRACE: + /* If it the next token is a `{', then we are entering a new + block. Consume the entire block. */ + ++nesting_depth; + break; + + default: + break; + } + + /* Consume the token. */ + c_parser_consume_token (parser); + } +} + /* Expect to be at the end of the pragma directive and consume an end of line marker. */ @@ -1582,7 +1627,7 @@ c_parser_skip_to_end_of_block_or_statement (c_parser *parser, here for secondary error recovery, after parser->error has been cleared. */ c_parser_consume_pragma (parser); - c_parser_skip_to_pragma_eol (parser); + c_parser_skip_to_pragma_eol (parser, false); parser->error = save_error; continue; @@ -28381,19 +28426,17 @@ c_parser_omp_context_selector (c_parser *parser, enum omp_tss_code set, case OMP_TRAIT_PROPERTY_DEV_NUM_EXPR: case OMP_TRAIT_PROPERTY_BOOL_EXPR: t = c_parser_expr_no_commas (parser, NULL).value; - if (t != error_mark_node) + if (t == error_mark_node) + return error_mark_node; + mark_exp_read (t); + t = c_fully_fold (t, false, NULL); + if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) { - mark_exp_read (t); - t = c_fully_fold (t, false, NULL); - if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) - error_at (token->location, - "property must be integer expression"); - else - properties = make_trait_property (NULL_TREE, t, - properties); + error_at (token->location, + "property must be integer expression"); + return error_mark_node; } - else - return error_mark_node; + properties = make_trait_property (NULL_TREE, t, properties); break; case OMP_TRAIT_PROPERTY_CLAUSE_LIST: if (sel == OMP_TRAIT_CONSTRUCT_SIMD) @@ -28495,11 +28538,14 @@ c_parser_omp_context_selector_specification (c_parser *parser, tree parms) tree selectors = c_parser_omp_context_selector (parser, set, parms); if (selectors == error_mark_node) - ret = error_mark_node; + { + c_parser_skip_to_closing_brace (parser); + ret = error_mark_node; + } else if (ret != error_mark_node) ret = make_trait_set_selector (set, selectors, ret); - braces.skip_until_found_close (parser); + braces.require_close (parser); if (c_parser_next_token_is (parser, CPP_COMMA)) c_parser_consume_token (parser); @@ -31148,7 +31194,6 @@ c_parser_omp_metadirective (c_parser *parser, bool *if_p) { error_at (match_loc, "too many %<otherwise%> or %<default%> " "clauses in %<metadirective%>"); - c_parser_skip_to_end_of_block_or_statement (parser, true); goto error; } default_seen = true; @@ -31157,14 +31202,12 @@ c_parser_omp_metadirective (c_parser *parser, bool *if_p) { error_at (match_loc, "%<otherwise%> or %<default%> clause " "must appear last in %<metadirective%>"); - c_parser_skip_to_end_of_block_or_statement (parser, true); goto error; } if (!default_p && strcmp (p, "when") != 0) { error_at (match_loc, "%qs is not valid for %qs", p, "metadirective"); - c_parser_skip_to_end_of_block_or_statement (parser, true); goto error; } @@ -31232,7 +31275,6 @@ c_parser_omp_metadirective (c_parser *parser, bool *if_p) if (i == 0) { error_at (loc, "expected directive name"); - c_parser_skip_to_end_of_block_or_statement (parser, true); goto error; } @@ -31300,7 +31342,10 @@ c_parser_omp_metadirective (c_parser *parser, bool *if_p) goto add; case CPP_CLOSE_PAREN: if (nesting_depth-- == 0) - break; + { + c_parser_consume_token (parser); + break; + } goto add; default: add: @@ -31312,8 +31357,6 @@ c_parser_omp_metadirective (c_parser *parser, bool *if_p) break; } - c_parser_consume_token (parser); - if (!skip) { c_token eol_token; @@ -31440,9 +31483,9 @@ c_parser_omp_metadirective (c_parser *parser, bool *if_p) return; error: + /* Skip the metadirective pragma. Do not skip the metadirective body. */ if (parser->in_pragma) - c_parser_skip_to_pragma_eol (parser); - c_parser_skip_to_end_of_block_or_statement (parser, true); + c_parser_skip_to_pragma_eol (parser, false); } /* Main entry point to parsing most OpenMP pragmas. */ diff --git a/gcc/config/aarch64/aarch64-cores.def b/gcc/config/aarch64/aarch64-cores.def index 1209630..24b7cd3 100644 --- a/gcc/config/aarch64/aarch64-cores.def +++ b/gcc/config/aarch64/aarch64-cores.def @@ -132,7 +132,7 @@ AARCH64_CORE("octeontx2f95mm", octeontx2f95mm, cortexa57, V8_2A, (CRYPTO, PROFI /* Fujitsu ('F') cores. */ AARCH64_CORE("a64fx", a64fx, a64fx, V8_2A, (F16, SVE), a64fx, 0x46, 0x001, -1) -AARCH64_CORE("fujitsu-monaka", fujitsu_monaka, cortexa57, V9_3A, (F16, FP8, LS64, RNG, CRYPTO, SVE2_AES, SVE2_BITPERM, SVE2_SHA3, SVE2_SM4), fujitsu_monaka, 0x46, 0x003, -1) +AARCH64_CORE("fujitsu-monaka", fujitsu_monaka, cortexa57, V9_3A, (F16, FAMINMAX, FP8FMA, FP8DOT2, FP8DOT4, LS64, LUT, RNG, CRYPTO, SVE2_AES, SVE2_BITPERM, SVE2_SHA3, SVE2_SM4), fujitsu_monaka, 0x46, 0x003, -1) /* HiSilicon ('H') cores. */ AARCH64_CORE("tsv110", tsv110, tsv110, V8_2A, (CRYPTO, F16), tsv110, 0x48, 0xd01, -1) diff --git a/gcc/cp/ChangeLog b/gcc/cp/ChangeLog index 59bc179..8983abf 100644 --- a/gcc/cp/ChangeLog +++ b/gcc/cp/ChangeLog @@ -1,3 +1,29 @@ +2025-05-30 Sandra Loosemore <sloosemore@baylibre.com> + + Backported from master: + 2025-05-29 Sandra Loosemore <sloosemore@baylibre.com> + + * parser.cc (cp_parser_omp_metadirective): Do not call + cp_parser_skip_to_end_of_block_or_statement after an error. + +2025-05-30 Sandra Loosemore <sloosemore@baylibre.com> + + Backported from master: + 2025-05-29 Sandra Loosemore <sloosemore@baylibre.com> + + PR c/120180 + * parser.cc (cp_parser_omp_metadirective): Only consume the + token if it is the expected close paren. + +2025-05-26 Tobias Burnus <tburnus@baylibre.com> + + Backported from master: + 2025-05-26 Tobias Burnus <tburnus@baylibre.com> + + PR c++/120413 + * semantics.cc (finish_omp_target_clauses_r): Handle + BIND_EXPR with empty BIND_EXPR_BLOCK. + 2025-05-23 Nathaniel Shead <nathanieloshead@gmail.com> Backported from master: diff --git a/gcc/cp/parser.cc b/gcc/cp/parser.cc index be86252..9c1d976 100644 --- a/gcc/cp/parser.cc +++ b/gcc/cp/parser.cc @@ -53390,7 +53390,6 @@ cp_parser_omp_metadirective (cp_parser *parser, cp_token *pragma_tok, { error_at (match_loc, "too many %<otherwise%> or %<default%> " "clauses in %<metadirective%>"); - cp_parser_skip_to_end_of_block_or_statement (parser, true); goto fail; } else @@ -53400,14 +53399,12 @@ cp_parser_omp_metadirective (cp_parser *parser, cp_token *pragma_tok, { error_at (match_loc, "%<otherwise%> or %<default%> clause " "must appear last in %<metadirective%>"); - cp_parser_skip_to_end_of_block_or_statement (parser, true); goto fail; } if (!default_p && strcmp (p, "when") != 0) { error_at (match_loc, "%qs is not valid for %qs", p, "metadirective"); - cp_parser_skip_to_end_of_block_or_statement (parser, true); goto fail; } @@ -53476,7 +53473,6 @@ cp_parser_omp_metadirective (cp_parser *parser, cp_token *pragma_tok, if (i == 0) { error_at (loc, "expected directive name"); - cp_parser_skip_to_end_of_block_or_statement (parser, true); goto fail; } @@ -53549,7 +53545,10 @@ cp_parser_omp_metadirective (cp_parser *parser, cp_token *pragma_tok, goto add; case CPP_CLOSE_PAREN: if (nesting_depth-- == 0) - break; + { + cp_lexer_consume_token (parser->lexer); + break; + } goto add; default: add: @@ -53561,8 +53560,6 @@ cp_parser_omp_metadirective (cp_parser *parser, cp_token *pragma_tok, break; } - cp_lexer_consume_token (parser->lexer); - if (!skip) { cp_token eol_token = {}; @@ -53694,11 +53691,8 @@ cp_parser_omp_metadirective (cp_parser *parser, cp_token *pragma_tok, return; fail: - /* Skip the metadirective pragma. */ + /* Skip the metadirective pragma. Do not skip the metadirective body. */ cp_parser_skip_to_pragma_eol (parser, pragma_tok); - - /* Skip the metadirective body. */ - cp_parser_skip_to_end_of_block_or_statement (parser, true); } diff --git a/gcc/doc/extend.texi b/gcc/doc/extend.texi index ec68c85..bad3408 100644 --- a/gcc/doc/extend.texi +++ b/gcc/doc/extend.texi @@ -7336,7 +7336,7 @@ truncate the copy without appending the terminating @code{NUL} character. Using the attribute makes it possible to suppress the warning. However, when the array is declared with the attribute the call to @code{strlen} is diagnosed because when the array doesn't contain a @code{NUL}-terminated -string the call is undefined. To copy, compare, of search non-string +string the call is undefined. To copy, compare, or search non-string character arrays use the @code{memcpy}, @code{memcmp}, @code{memchr}, and other functions that operate on arrays of bytes. In addition, calling @code{strnlen} and @code{strndup} with such arrays is safe diff --git a/gcc/fortran/ChangeLog b/gcc/fortran/ChangeLog index b95f7ee..2f9f5c9 100644 --- a/gcc/fortran/ChangeLog +++ b/gcc/fortran/ChangeLog @@ -1,3 +1,28 @@ +2025-05-31 Jerry DeLisle <jvdelisle@gcc.gnu.org> + + Backported from master: + 2025-05-27 Jerry DeLisle <jvdelisle@gcc.gnu.org> + + PR fortran/120049 + * check.cc (gfc_check_c_associated): Use new helper functions. + Only call check_c_ptr_1 if optional c_ptr_2 tests succeed. + (check_c_ptr_1): Handle only c_ptr_1 checks. + (check_c_ptr_2): Expand checks for c_ptr_2 and handle cases + where there is no derived pointer in the gfc_expr and check + the inmod_sym_id only if it exists. Rephrase error message. + * misc.cc (gfc_typename): Handle the case for BT_VOID rather + than throw an internal error. + +2025-05-31 Thomas Koenig <tkoenig@gcc.gnu.org> + + Backported from master: + 2025-05-30 Thomas Koenig <tkoenig@gcc.gnu.org> + + PR fortran/120355 + * interface.cc (compare_parameter): If the global function has a + result clause, take typespec from there for the comparison against + the dummy argument. + 2025-05-22 Harald Anlauf <anlauf@gmx.de> Backported from master: diff --git a/gcc/fortran/check.cc b/gcc/fortran/check.cc index 0073cd0..ce9da31 100644 --- a/gcc/fortran/check.cc +++ b/gcc/fortran/check.cc @@ -5916,49 +5916,110 @@ gfc_check_c_sizeof (gfc_expr *arg) } -bool -gfc_check_c_associated (gfc_expr *c_ptr_1, gfc_expr *c_ptr_2) +/* Helper functions check_c_ptr_1 and check_c_ptr_2 + used in gfc_check_c_associated. */ + +static inline +bool check_c_ptr_1 (gfc_expr *c_ptr_1) { - if (c_ptr_1) - { - if (c_ptr_1->expr_type == EXPR_FUNCTION && c_ptr_1->ts.type == BT_VOID) - return true; + if ((c_ptr_1->ts.type == BT_VOID) + && (c_ptr_1->expr_type == EXPR_FUNCTION)) + return true; - if (c_ptr_1->ts.type != BT_DERIVED - || c_ptr_1->ts.u.derived->from_intmod != INTMOD_ISO_C_BINDING - || (c_ptr_1->ts.u.derived->intmod_sym_id != ISOCBINDING_PTR - && c_ptr_1->ts.u.derived->intmod_sym_id != ISOCBINDING_FUNPTR)) - { - gfc_error ("Argument C_PTR_1 at %L to C_ASSOCIATED shall have the " - "type TYPE(C_PTR) or TYPE(C_FUNPTR)", &c_ptr_1->where); - return false; - } - } + if (c_ptr_1->ts.type != BT_DERIVED + || c_ptr_1->ts.u.derived->from_intmod != INTMOD_ISO_C_BINDING + || (c_ptr_1->ts.u.derived->intmod_sym_id != ISOCBINDING_PTR + && c_ptr_1->ts.u.derived->intmod_sym_id != ISOCBINDING_FUNPTR)) + goto check_1_error; - if (!scalar_check (c_ptr_1, 0)) + if ((c_ptr_1->ts.type == BT_DERIVED) + && (c_ptr_1->expr_type == EXPR_STRUCTURE) + && (c_ptr_1->ts.u.derived->intmod_sym_id + == ISOCBINDING_NULL_FUNPTR)) + goto check_1_error; + + if (scalar_check (c_ptr_1, 0)) + return true; + else + /* Return since the check_1_error message may not apply here. */ return false; - if (c_ptr_2) - { - if (c_ptr_2->expr_type == EXPR_FUNCTION && c_ptr_2->ts.type == BT_VOID) - return true; +check_1_error: - if (c_ptr_2->ts.type != BT_DERIVED - || c_ptr_2->ts.u.derived->from_intmod != INTMOD_ISO_C_BINDING - || (c_ptr_1->ts.u.derived->intmod_sym_id - != c_ptr_2->ts.u.derived->intmod_sym_id)) + gfc_error ("Argument C_PTR_1 at %L to C_ASSOCIATED shall have the " + "type TYPE(C_PTR) or TYPE(C_FUNPTR)", &c_ptr_1->where); + return false; +} + +static inline +bool check_c_ptr_2 (gfc_expr *c_ptr_1, gfc_expr *c_ptr_2) +{ + switch (c_ptr_2->ts.type) + { + case BT_VOID: + if (c_ptr_2->expr_type == EXPR_FUNCTION) { - gfc_error ("Argument C_PTR_2 at %L to C_ASSOCIATED shall have the " - "same type as C_PTR_1: %s instead of %s", &c_ptr_1->where, - gfc_typename (&c_ptr_1->ts), gfc_typename (&c_ptr_2->ts)); - return false; + if ((c_ptr_1->ts.type == BT_DERIVED) + && c_ptr_1->expr_type == EXPR_STRUCTURE + && (c_ptr_1->ts.u.derived->intmod_sym_id + == ISOCBINDING_FUNPTR)) + goto check_2_error; } - } + break; + + case BT_DERIVED: + if ((c_ptr_2->expr_type == EXPR_STRUCTURE) + && (c_ptr_2->ts.u.derived->intmod_sym_id == ISOCBINDING_PTR) + && (c_ptr_1->ts.type == BT_VOID) + && (c_ptr_1->expr_type == EXPR_FUNCTION)) + return scalar_check (c_ptr_2, 1); + + if ((c_ptr_2->expr_type == EXPR_STRUCTURE) + && (c_ptr_1->ts.type == BT_VOID) + && (c_ptr_1->expr_type == EXPR_FUNCTION)) + goto check_2_error; - if (c_ptr_2 && !scalar_check (c_ptr_2, 1)) + if (c_ptr_2->ts.u.derived->from_intmod != INTMOD_ISO_C_BINDING) + goto check_2_error; + + if (c_ptr_1->ts.type == BT_DERIVED + && (c_ptr_1->ts.u.derived->intmod_sym_id + != c_ptr_2->ts.u.derived->intmod_sym_id)) + goto check_2_error; + break; + + default: + goto check_2_error; + } + + if (scalar_check (c_ptr_2, 1)) + return true; + else + /* Return since the check_2_error message may not apply here. */ return false; - return true; +check_2_error: + + gfc_error ("Argument C_PTR_2 at %L to C_ASSOCIATED shall have the " + "same type as C_PTR_1, found %s instead of %s", &c_ptr_2->where, + gfc_typename (&c_ptr_2->ts), gfc_typename (&c_ptr_1->ts)); + + return false; + } + + +bool +gfc_check_c_associated (gfc_expr *c_ptr_1, gfc_expr *c_ptr_2) +{ + if (c_ptr_2) + { + if (check_c_ptr_2 (c_ptr_1, c_ptr_2)) + return check_c_ptr_1 (c_ptr_1); + else + return false; + } + else + return check_c_ptr_1 (c_ptr_1); } diff --git a/gcc/fortran/interface.cc b/gcc/fortran/interface.cc index 753f589..b854292 100644 --- a/gcc/fortran/interface.cc +++ b/gcc/fortran/interface.cc @@ -2547,7 +2547,14 @@ compare_parameter (gfc_symbol *formal, gfc_expr *actual, } else if (formal->attr.function) { - if (!gfc_compare_types (&global_asym->ts, + gfc_typespec ts; + + if (global_asym->result) + ts = global_asym->result->ts; + else + ts = global_asym->ts; + + if (!gfc_compare_types (&ts, &formal->ts)) { gfc_error ("Type mismatch at %L passing global " diff --git a/gcc/fortran/io.cc b/gcc/fortran/io.cc index b5c9d33..7466d8f 100644 --- a/gcc/fortran/io.cc +++ b/gcc/fortran/io.cc @@ -1228,7 +1228,8 @@ between_desc: default: if (mode != MODE_FORMAT) format_locus.nextc += format_string_pos - 1; - if (!gfc_notify_std (GFC_STD_GNU, "Missing comma at %L", &format_locus)) + if (!gfc_notify_std (GFC_STD_LEGACY, + "Missing comma in FORMAT string at %L", &format_locus)) return false; /* If we do not actually return a failure, we need to unwind this before the next round. */ @@ -1290,7 +1291,8 @@ extension_optional_comma: default: if (mode != MODE_FORMAT) format_locus.nextc += format_string_pos; - if (!gfc_notify_std (GFC_STD_GNU, "Missing comma at %L", &format_locus)) + if (!gfc_notify_std (GFC_STD_LEGACY, + "Missing comma in FORMAT string at %L", &format_locus)) return false; /* If we do not actually return a failure, we need to unwind this before the next round. */ diff --git a/gcc/fortran/misc.cc b/gcc/fortran/misc.cc index 893c40f..b8bdf75 100644 --- a/gcc/fortran/misc.cc +++ b/gcc/fortran/misc.cc @@ -214,6 +214,9 @@ gfc_typename (gfc_typespec *ts, bool for_hash) case BT_UNKNOWN: strcpy (buffer, "UNKNOWN"); break; + case BT_VOID: + strcpy (buffer, "VOID"); + break; default: gfc_internal_error ("gfc_typename(): Undefined type"); } diff --git a/gcc/ipa-sra.cc b/gcc/ipa-sra.cc index 88bfae9..6e6cf89 100644 --- a/gcc/ipa-sra.cc +++ b/gcc/ipa-sra.cc @@ -1848,6 +1848,12 @@ scan_expr_access (tree expr, gimple *stmt, isra_scan_context ctx, if (!desc || !desc->split_candidate) return; + if (storage_order_barrier_p (expr)) + { + disqualify_split_candidate (desc, "Encountered a storage order barrier."); + return; + } + if (!poffset.is_constant (&offset) || !psize.is_constant (&size) || !pmax_size.is_constant (&max_size)) diff --git a/gcc/testsuite/ChangeLog b/gcc/testsuite/ChangeLog index 750a1b9..5b94c72 100644 --- a/gcc/testsuite/ChangeLog +++ b/gcc/testsuite/ChangeLog @@ -1,3 +1,75 @@ +2025-05-31 Jerry DeLisle <jvdelisle@gcc.gnu.org> + + Backported from master: + 2025-05-27 Jerry DeLisle <jvdelisle@gcc.gnu.org> + + PR fortran/120049 + * gfortran.dg/pr120049_a.f90: Update test directives. + * gfortran.dg/pr120049_b.f90: Update test directives + * gfortran.dg/pr120049_2.f90: New test. + * gfortran.dg/c_f_pointer_tests_6.f90: Adjust dg-error + directive. + Co-Authored-By: Steve Kargl <kargl@gcc.gnu.org> + +2025-05-31 Thomas Koenig <tkoenig@gcc.gnu.org> + + Backported from master: + 2025-05-30 Thomas Koenig <tkoenig@gcc.gnu.org> + + PR fortran/120355 + * gfortran.dg/interface_62.f90: New test. + +2025-05-30 Sandra Loosemore <sloosemore@baylibre.com> + + Backported from master: + 2025-05-29 Sandra Loosemore <sloosemore@baylibre.com> + + * c-c++-common/gomp/declare-variant-2.c: Adjust patterns now that + C and C++ now behave similarly. + * c-c++-common/gomp/metadirective-error-recovery.c: New. + +2025-05-30 Sandra Loosemore <sloosemore@baylibre.com> + + Backported from master: + 2025-05-29 Sandra Loosemore <sloosemore@baylibre.com> + + PR c/120180 + * c-c++-common/gomp/pr120180.c: New. + +2025-05-30 Jakub Jelinek <jakub@redhat.com> + + Backported from master: + 2025-05-30 Jakub Jelinek <jakub@redhat.com> + + PR target/120480 + * gcc.dg/pr120480.c: New test. + +2025-05-27 Eric Botcazou <ebotcazou@adacore.com> + + * gnat.dg/sso19.adb: New test. + * gnat.dg/sso19_pkg.ads, gnat.dg/sso19_pkg.adb: New helper. + +2025-05-26 Tobias Burnus <tburnus@baylibre.com> + + Backported from master: + 2025-05-26 Tobias Burnus <tburnus@baylibre.com> + + PR middle-end/118694 + * c-c++-common/gomp/attrs-metadirective-3.c: Change to never + expect 'omp metadirective' in the dump. If !offload_nvptx, check + that no 'teams' shows up in the dump; for offload_nvptx, expect + OMP_NEXT_VARIANT and an error about directive between 'target' + and 'teams'. + * c-c++-common/gomp/metadirective-3.c: Likewise. + +2025-05-26 Tobias Burnus <tburnus@baylibre.com> + + Backported from master: + 2025-05-26 Tobias Burnus <tburnus@baylibre.com> + + PR c++/120413 + * g++.dg/gomp/target-4.C: New test. + 2025-05-23 Nathaniel Shead <nathanieloshead@gmail.com> Backported from master: diff --git a/gcc/testsuite/c-c++-common/gomp/declare-variant-2.c b/gcc/testsuite/c-c++-common/gomp/declare-variant-2.c index 7711dbc..f8f5143 100644 --- a/gcc/testsuite/c-c++-common/gomp/declare-variant-2.c +++ b/gcc/testsuite/c-c++-common/gomp/declare-variant-2.c @@ -47,10 +47,9 @@ void f23 (void); #pragma omp declare variant (f1) match(construct={teams,parallel,master,for}) /* { dg-warning "unknown selector 'master' for context selector set 'construct'" } */ void f24 (void); #pragma omp declare variant (f1) match(construct={parallel(1 /* { dg-error "selector 'parallel' does not accept any properties" } */ -void f25 (void); /* { dg-error "expected '\\\}' before end of line" "" { target c++ } .-1 } */ - /* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-2 } */ +void f25 (void); /* { dg-error "expected '\\\}' before end of line" "" { target *-*-* } .-1 } */ #pragma omp declare variant (f1) match(construct={parallel(1)}) /* { dg-error "selector 'parallel' does not accept any properties" } */ -void f26 (void); /* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */ +void f26 (void); #pragma omp declare variant (f0) match(construct={simd(12)}) /* { dg-error "expected \[^\n\r]* clause before" } */ void f27 (void); /* { dg-error "'\\)' before numeric constant" "" { target c++ } .-1 } */ #pragma omp declare variant (f1) match(construct={parallel},construct={for}) /* { dg-error "selector set 'construct' specified more than once" } */ @@ -96,13 +95,13 @@ void f46 (void); #pragma omp declare variant (f1) match(implementation={vendor("foobar")}) /* { dg-warning "unknown property '.foobar.' of 'vendor' selector" } */ void f47 (void); #pragma omp declare variant (f1) match(implementation={unified_address(yes)}) /* { dg-error "selector 'unified_address' does not accept any properties" } */ -void f48 (void); /* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */ +void f48 (void); #pragma omp declare variant (f1) match(implementation={unified_shared_memory(no)}) /* { dg-error "selector 'unified_shared_memory' does not accept any properties" } */ -void f49 (void); /* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */ +void f49 (void); #pragma omp declare variant (f1) match(implementation={dynamic_allocators(42)}) /* { dg-error "selector 'dynamic_allocators' does not accept any properties" } */ -void f50 (void); /* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */ +void f50 (void); #pragma omp declare variant (f1) match(implementation={reverse_offload()}) /* { dg-error "selector 'reverse_offload' does not accept any properties" } */ -void f51 (void); /* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */ +void f51 (void); #pragma omp declare variant (f1) match(implementation={atomic_default_mem_order}) /* { dg-error "expected '\\(' before '\\\}' token" } */ void f52 (void); #pragma omp declare variant (f1) match(implementation={atomic_default_mem_order(acquire)}) diff --git a/gcc/testsuite/c-c++-common/gomp/metadirective-error-recovery.c b/gcc/testsuite/c-c++-common/gomp/metadirective-error-recovery.c new file mode 100644 index 0000000..3242281 --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/metadirective-error-recovery.c @@ -0,0 +1,20 @@ +/* { dg-do compile } */ + +/* This test used to ICE in C and only diagnose the first error in C++. */ + +struct s { + int a, b; +}; + +void f (int aa, int bb) +{ + struct s s1, s2; + s1.a = aa; + s1.b = bb; + s2.a = aa + 1; + s2.b = bb + 1; + + /* A struct is not a valid argument for the condition selector. */ + #pragma omp metadirective when(user={condition(s1)} : nothing) otherwise(nothing) /* { dg-error "property must be integer expression" } */ + #pragma omp metadirective when(user={condition(s2)} : nothing) otherwise(nothing) /* { dg-error "property must be integer expression" } */ +} diff --git a/gcc/testsuite/c-c++-common/gomp/pr120180.c b/gcc/testsuite/c-c++-common/gomp/pr120180.c new file mode 100644 index 0000000..cb5a0d5 --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/pr120180.c @@ -0,0 +1,22 @@ +/* { dg-do compile } */ + +/* This test used to ICE after erroring on the metadirective in the + loop nest. */ + +int main() +{ + int blksize = 15000; + double *qq; + int i, k, nq; + + #pragma omp metadirective when(user={condition(0)}: target teams distribute parallel for collapse(2) map(qq[:0]) private(i)) \ + when(user={condition(0)}: target teams distribute parallel for map(qq[:0]) private(i)) \ + when(user={condition(1)}: target teams loop collapse(2) map(qq[:0]) private(i)) + for(k=0; k<blksize; k++) + { +#pragma omp metadirective when(user={condition(0)}: simd) default() // { dg-error "intervening code must not contain OpenMP directives" } + for (i=0; i<nq; i++) + qq[k*nq + i] = 0.0; + } + return 0; +} diff --git a/gcc/testsuite/gcc.dg/pr120480.c b/gcc/testsuite/gcc.dg/pr120480.c new file mode 100644 index 0000000..cf7b47a --- /dev/null +++ b/gcc/testsuite/gcc.dg/pr120480.c @@ -0,0 +1,11 @@ +/* PR target/120480 */ +/* { dg-do compile } */ +/* { dg-options "-O0" } */ + +struct S { int a, b, c; } s; + +void +foo (void) +{ + struct S t = s; +} diff --git a/gcc/testsuite/gfortran.dg/c_f_pointer_tests_6.f90 b/gcc/testsuite/gfortran.dg/c_f_pointer_tests_6.f90 index 23ca88b..bc2206d 100644 --- a/gcc/testsuite/gfortran.dg/c_f_pointer_tests_6.f90 +++ b/gcc/testsuite/gfortran.dg/c_f_pointer_tests_6.f90 @@ -38,6 +38,6 @@ contains type(my_c_ptr_0) :: my_ptr2 type(c_funptr) :: myfun print *,c_associated(my_ptr,my_ptr2) - print *,c_associated(my_ptr,myfun) ! { dg-error "Argument C_PTR_2 at .1. to C_ASSOCIATED shall have the same type as C_PTR_1: TYPE.c_ptr. instead of TYPE.c_funptr." } + print *,c_associated(my_ptr,myfun) ! { dg-error "Argument C_PTR_2 at .1. to C_ASSOCIATED shall have the same type as C_PTR_1, found TYPE.c_funptr. instead of TYPE.c_ptr." } end subroutine end diff --git a/gcc/testsuite/gfortran.dg/comma_format_extension_1.f b/gcc/testsuite/gfortran.dg/comma_format_extension_1.f index a3a5a98..c4b43f0 100644 --- a/gcc/testsuite/gfortran.dg/comma_format_extension_1.f +++ b/gcc/testsuite/gfortran.dg/comma_format_extension_1.f @@ -1,5 +1,5 @@ ! { dg-do compile } -! { dg-options "" } +! { dg-options "-std=legacy" } ! test that the extension for a missing comma is accepted subroutine mysub diff --git a/gcc/testsuite/gfortran.dg/comma_format_extension_3.f b/gcc/testsuite/gfortran.dg/comma_format_extension_3.f index 0b00224..9d974d6 100644 --- a/gcc/testsuite/gfortran.dg/comma_format_extension_3.f +++ b/gcc/testsuite/gfortran.dg/comma_format_extension_3.f @@ -3,7 +3,7 @@ ! did do the correct thing at runtime. ! Note the missing , before i1 in the format. ! { dg-do run } -! { dg-options "" } +! { dg-options "-std=legacy" } character*12 c write (c,100) 0, 1 diff --git a/gcc/testsuite/gfortran.dg/continuation_13.f90 b/gcc/testsuite/gfortran.dg/continuation_13.f90 index 9799b59e..475c896 100644 --- a/gcc/testsuite/gfortran.dg/continuation_13.f90 +++ b/gcc/testsuite/gfortran.dg/continuation_13.f90 @@ -1,5 +1,5 @@ ! { dg-do run } -! { dg-options "-std=gnu" } +! { dg-options "-std=legacy" } ! PR64506 character(25) :: astring diff --git a/gcc/testsuite/gfortran.dg/interface_62.f90 b/gcc/testsuite/gfortran.dg/interface_62.f90 new file mode 100644 index 0000000..19d4325 --- /dev/null +++ b/gcc/testsuite/gfortran.dg/interface_62.f90 @@ -0,0 +1,39 @@ +! { dg-do compile } +! PR fortran/120355 - this was rejected because the typespec from +! the RESULT clause was not picked up. +! Test case jsberg@bnl.gov. + +program p + implicit none + integer :: i,j + interface + function s(x) result(y) + implicit none + integer, intent(in) :: x + integer :: y + end function s + end interface + i = 0 + call t(s,i,j) +contains + subroutine t(f,x,y) + implicit none + integer, intent(in) :: x + integer, intent(out) :: y + interface + function f(x) result(y) + implicit none + integer, intent(in) :: x + integer :: y + end function f + end interface + y = f(x) + end subroutine t +end program p + +function s(x) result(y) + implicit none + integer, intent(in) :: x + integer :: y + y = 1 - x +end function s diff --git a/gcc/testsuite/gfortran.dg/pr119856.f90 b/gcc/testsuite/gfortran.dg/pr119856.f90 new file mode 100644 index 0000000..60ada0a --- /dev/null +++ b/gcc/testsuite/gfortran.dg/pr119856.f90 @@ -0,0 +1,15 @@ +! { dg-do run } +! PR119856, the error should occur in both write statements. +program badfmt + implicit none + + character(10):: fmt = "(AI5)" ! Not a PARAMETER so not examined + ! at compile time + integer :: ioerr + ioerr = 0 + write (*, fmt, iostat=ioerr) 'value =', 42 + if (ioerr /= 5006) stop 10 +! + write (*, fmt, iostat=ioerr) 'value =', 43 + if (ioerr /= 5006) stop 13 +end program badfmt diff --git a/gcc/testsuite/gfortran.dg/pr120049_2.f90 b/gcc/testsuite/gfortran.dg/pr120049_2.f90 new file mode 100644 index 0000000..1f91e06 --- /dev/null +++ b/gcc/testsuite/gfortran.dg/pr120049_2.f90 @@ -0,0 +1,62 @@ +! Compiled with pr120049_b.f90 +! { dg-options -O0 } +! { dg-do compile } +! { dg-compile-aux-modules "pr120049_b.f90" } +! +! Test the fix for PR120049 +program tests_gtk_sup + use gtk_sup + implicit none + + type mytype + integer :: myint + end type mytype + type(mytype) :: ijkl = mytype(42) + logical :: truth + real :: var1 + type(c_ptr), target :: val + type(c_funptr), target :: fptr + character(15) :: stringy + complex :: certainly + truth = .true. + var1 = 86. + stringy = "what the hay!" + certainly = (3.14,-4.13) + if (c_associated(val, c_loc(val))) then + stop 1 + endif + if (c_associated(c_loc(val), val)) then + stop 2 + endif + print *, c_associated(fptr, C_NULL_FUNPTR) + print *, c_associated(c_loc(val), C_NULL_PTR) + print *, c_associated(C_NULL_PTR, c_loc(val)) + print *, c_associated(c_loc(val), 42) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(c_loc(val), .42) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(c_loc(val), truth) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(c_loc(val), .false.) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(c_loc(val), var1) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(c_loc(val), stringy) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(c_loc(val), certainly) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(42) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(.42) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(truth) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(.false.) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(var1) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(stringy) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(certainly) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(.42) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(val, testit(val)) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(testit(val), val) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(testit(val)) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(c_loc(val), C_NULL_FUNPTR) ! { dg-error "C_ASSOCIATED shall have the" } + print *, c_associated(C_NULL_FUNPTR, c_loc(val)) ! { dg-error "C_ASSOCIATED shall have the" } +contains + + function testit (avalue) result(res) + type(c_ptr) :: avalue + type(mytype) :: res + res%myint = 42 + end function + +end program tests_gtk_sup diff --git a/gcc/testsuite/gfortran.dg/pr120049_a.f90 b/gcc/testsuite/gfortran.dg/pr120049_a.f90 index c404a4d..7095314 100644 --- a/gcc/testsuite/gfortran.dg/pr120049_a.f90 +++ b/gcc/testsuite/gfortran.dg/pr120049_a.f90 @@ -1,5 +1,8 @@ -! { dg-do preprocess } -! { dg-additional-options "-cpp" } +! Compiled with pr120049_b.f90 +! { dg-options -O0 } +! { dg-do run } +! { dg-compile-aux-modules "pr120049_b.f90" } +! { dg-additional-sources pr120049_b.f90 } ! ! Test the fix for PR86248 program tests_gtk_sup diff --git a/gcc/testsuite/gfortran.dg/pr120049_b.f90 b/gcc/testsuite/gfortran.dg/pr120049_b.f90 index 127db98..28a2783 100644 --- a/gcc/testsuite/gfortran.dg/pr120049_b.f90 +++ b/gcc/testsuite/gfortran.dg/pr120049_b.f90 @@ -1,5 +1,3 @@ -! { dg-do run } -! { dg-additional-sources pr120049_a.f90 } ! ! Module for pr120049.f90 ! diff --git a/gcc/testsuite/gnat.dg/sso19.adb b/gcc/testsuite/gnat.dg/sso19.adb new file mode 100644 index 0000000..497d9874 --- /dev/null +++ b/gcc/testsuite/gnat.dg/sso19.adb @@ -0,0 +1,13 @@ +-- { dg-do run } +-- { dg-options "-O2" } + +with SSO19_Pkg; use SSO19_Pkg; + +procedure SSO19 is + R : constant Rec := (D => (I => 8, F => 4.6095713E-41)); + +begin + if not Is_Valid (R) then + raise Program_Error; + end if; +end; diff --git a/gcc/testsuite/gnat.dg/sso19_pkg.adb b/gcc/testsuite/gnat.dg/sso19_pkg.adb new file mode 100644 index 0000000..cbcb2f9 --- /dev/null +++ b/gcc/testsuite/gnat.dg/sso19_pkg.adb @@ -0,0 +1,13 @@ +package body SSO19_Pkg is + + function Is_Valid_Private (Item : Data) return Boolean is + begin + return Item.I'Valid and Item.F'Valid; + end Is_Valid_Private; + + function Is_Valid (Item : Rec) return Boolean is + begin + return Is_Valid_Private (Item.D); + end Is_Valid; + +end SSO19_Pkg; diff --git a/gcc/testsuite/gnat.dg/sso19_pkg.ads b/gcc/testsuite/gnat.dg/sso19_pkg.ads new file mode 100644 index 0000000..cad5368 --- /dev/null +++ b/gcc/testsuite/gnat.dg/sso19_pkg.ads @@ -0,0 +1,24 @@ +with System; + +package SSO19_Pkg is + + subtype Small_Int is Short_Integer range -1000 .. 1000; + + type Data is record + I : Small_Int; + F : Float; + end record; + for Data use record + I at 0 range 0 .. 15; + F at 4 range 0 .. 31; + end record; + for Data'Bit_Order use System.High_Order_First; + for Data'Scalar_Storage_Order use System.High_Order_First; + + type Rec is record + D : Data; + end record; + + function Is_Valid (Item : Rec) return Boolean; + +end SSO19_Pkg; diff --git a/gcc/tree-nrv.cc b/gcc/tree-nrv.cc index 180ce39..3be97af 100644 --- a/gcc/tree-nrv.cc +++ b/gcc/tree-nrv.cc @@ -167,16 +167,21 @@ pass_nrv::execute (function *fun) for (gsi = gsi_start_bb (bb); !gsi_end_p (gsi); gsi_next (&gsi)) { gimple *stmt = gsi_stmt (gsi); - tree ret_val; if (greturn *return_stmt = dyn_cast <greturn *> (stmt)) { - /* In a function with an aggregate return value, the - gimplifier has changed all non-empty RETURN_EXPRs to - return the RESULT_DECL. */ - ret_val = gimple_return_retval (return_stmt); - if (ret_val) - gcc_assert (ret_val == result); + /* We cannot perform NRV optimizations in a function with an + aggregate return value if there is a return that does not + return RESULT_DECL. We used to assert this scenario doesn't + happen: the gimplifier has changed all non-empty RETURN_EXPRs + to return the RESULT_DECL. However, per PR119835 we may run + into this scenario for offloading compilation, and therefore + gracefully bail out. */ + if (tree ret_val = gimple_return_retval (return_stmt)) + { + if (ret_val != result) + return 0; + } } else if (gimple_has_lhs (stmt) && gimple_get_lhs (stmt) == result) @@ -5572,7 +5572,7 @@ storage_order_barrier_p (const_tree t) && TYPE_REVERSE_STORAGE_ORDER (TREE_TYPE (op))) return true; - return false; + return reverse_storage_order_for_component_p (op); } /* Given a DECL or TYPE, return the scope in which it was declared, or diff --git a/libgcc/config/avr/libf7/ChangeLog b/libgcc/config/avr/libf7/ChangeLog index d7637ee..0e6afa0 100644 --- a/libgcc/config/avr/libf7/ChangeLog +++ b/libgcc/config/avr/libf7/ChangeLog @@ -1,3 +1,24 @@ +2025-05-27 Georg-Johann Lay <avr@gjlay.de> + + Backported from master: + 2025-05-27 Georg-Johann Lay <avr@gjlay.de> + + PR target/120442 + * libf7-common.mk (LIBF_C_PARTS, m_ddd): Add fdim. + * libf7.h (f7_fdim): New proto. + * libf7.c (f7_fdim): New function. + * f7renames.sh (f7_fdim): Add rename. + * f7-wraps.h: Rebuild + * f7-renames.h: Rebuild + +2025-05-27 Georg-Johann Lay <avr@gjlay.de> + + Backported from master: + 2025-05-27 Georg-Johann Lay <avr@gjlay.de> + + PR target/120441 + * libf7.c (f7_exp): Limit aa->expo to 10 (not to 9). + 2025-04-25 Release Manager * GCC 15.1.0 released. diff --git a/libgcc/config/avr/libf7/f7-renames.h b/libgcc/config/avr/libf7/f7-renames.h index bbe571a..bce2dd3 100644 --- a/libgcc/config/avr/libf7/f7-renames.h +++ b/libgcc/config/avr/libf7/f7-renames.h @@ -97,6 +97,7 @@ #define f7_acos __f7_acos #define f7_atan __f7_atan #define f7_atan2 __f7_atan2 +#define f7_fdim __f7_fdim #define f7_mul_noround __f7_mul_noround #define f7_sqrt16_round __f7_sqrt16_round #define f7_sqrt16_floor __f7_sqrt16_floor diff --git a/libgcc/config/avr/libf7/f7-wraps.h b/libgcc/config/avr/libf7/f7-wraps.h index a455b7d..409492e 100644 --- a/libgcc/config/avr/libf7/f7-wraps.h +++ b/libgcc/config/avr/libf7/f7-wraps.h @@ -239,7 +239,7 @@ _ENDF __extendsfdf2 ;; Functions that usually live in libm: Depending on [long] double layout, ;; define <name> and <name>l as weak alias(es) of __<name> for <name> in: -;; pow fmin fmax fmod hypot atan2 +;; pow fmin fmax fmod hypot atan2 fdim ;; double __pow (double, double) #ifdef F7MOD_D_pow_ @@ -313,6 +313,18 @@ _DEFUN __atan2 _ENDF __atan2 #endif /* F7MOD_D_atan2_ */ +;; double __fdim (double, double) +#ifdef F7MOD_D_fdim_ +_DEFUN __fdim + DALIAS fdim + LALIAS fdiml + .global F7_NAME(fdim) + ldi ZH, hi8(gs(F7_NAME(fdim))) + ldi ZL, lo8(gs(F7_NAME(fdim))) + F7jmp call_ddd +_ENDF __fdim +#endif /* F7MOD_D_fdim_ */ + ;; Functions that usually live in libm: Depending on [long] double layout, ;; define <name> and <name>l as weak alias(es) of __<name> for <name> in: ;; ldexp frexp diff --git a/libgcc/config/avr/libf7/f7renames.sh b/libgcc/config/avr/libf7/f7renames.sh index 7ef251e..4ced423 100755 --- a/libgcc/config/avr/libf7/f7renames.sh +++ b/libgcc/config/avr/libf7/f7renames.sh @@ -35,9 +35,9 @@ EOF c) if [ x${PRE} != xf7_ ]; then - echo " " + echo "" echo "/* Renames for libf7.c, libf7.h. */" - echo " " + echo "" for x in $*; do echo "#define f7_$x ${PRE}$x" done @@ -46,9 +46,9 @@ EOF cst) if [ x${PRE} != xf7_ ]; then - echo " " + echo "" echo "/* Renames for libf7.c, libf7.h. */" - echo " " + echo "" for x in $*; do echo "#define f7_const_${x} ${PRE}const_${x}" echo "#define f7_const_${x}_P ${PRE}const_${x}_P" @@ -58,9 +58,9 @@ EOF asm) if [ x${PRE} != xf7_ ]; then - echo " " + echo "" echo "/* Renames for libf7-asm.sx, f7-wraps.h. */" - echo " " + echo "" for x in $*; do echo "#define f7_${x}_asm ${PRE}${x}_asm" done diff --git a/libgcc/config/avr/libf7/libf7-common.mk b/libgcc/config/avr/libf7/libf7-common.mk index 5d41107..644be2c 100644 --- a/libgcc/config/avr/libf7/libf7-common.mk +++ b/libgcc/config/avr/libf7/libf7-common.mk @@ -8,7 +8,7 @@ F7_C_PARTS += set_float get_float get_double set_double set_pdouble F7_C_PARTS += fabs neg fmin fmax minmax truncx trunc floor ceil round lround F7_C_PARTS += horner logx log log10 log2 exp pow10 pow powi F7_C_PARTS += sin cos tan cotan sincos sinh cosh tanh sinhcosh -F7_C_PARTS += asinacos asin acos atan atan2 +F7_C_PARTS += asinacos asin acos atan atan2 fdim F7_C_PARTS += abscmp_msb_ge cmp cmp_abs cmp_unordered F7_C_PARTS += const_1 const_1_2 const_1_3 @@ -34,7 +34,7 @@ g_xdd_cmp += le lt ge gt ne eq unord g_dx += floatunsidf floatsidf extendsfdf2 g_xd += fixdfsi fixdfdi fixunsdfdi fixunsdfsi truncdfsf2 -m_ddd += pow fmin fmax fmod hypot atan2 +m_ddd += pow fmin fmax fmod hypot atan2 fdim m_ddx += ldexp frexp m_dd += sqrt cbrt exp exp10 pow10 log log10 log2 sin cos tan cotan asin acos atan m_dd += ceil floor trunc round sinh cosh tanh @@ -59,7 +59,7 @@ F7F += lrint ldexp frexp exp logx log log10 log2 F7F += minmax fmax fmin floor ceil round lround trunc truncx F7F += horner pow10 exp10 pow powi F7F += sin cos tan cotan sincos sinh cosh tanh sinhcosh -F7F += asinacos asin acos atan atan2 +F7F += asinacos asin acos atan atan2 fdim F7F += mul_noround sqrt16_round sqrt16_floor F7F += clr_mant_lsbs abscmp_msb_ge lshrdi3 ashldi3 F7F += assert diff --git a/libgcc/config/avr/libf7/libf7.c b/libgcc/config/avr/libf7/libf7.c index a64554c..78c218a 100644 --- a/libgcc/config/avr/libf7/libf7.c +++ b/libgcc/config/avr/libf7/libf7.c @@ -928,6 +928,21 @@ void f7_sub (f7_t *cc, const f7_t *aa, const f7_t *bb) #endif // F7MOD_sub_ +#ifdef F7MOD_fdim_ +F7_WEAK +void f7_fdim (f7_t *cc, const f7_t *aa, const f7_t *bb) +{ + int8_t cmp = f7_cmp_unordered (aa, bb, true /*with_sign*/); + if (cmp == INT8_MIN) + return f7_set_nan (cc); + if (cmp < 0) + return f7_clr (cc); + + f7_sub (cc, aa, bb); +} +#endif // F7MOD_fdim_ + + #ifdef F7MOD_addsub_ static void return_with_sign (f7_t *cc, const f7_t *aa, int8_t c_sign) { @@ -1649,10 +1664,10 @@ void f7_exp (f7_t *cc, const f7_t *aa) return f7_set_nan (cc); /* The maximal exponent of 2 for a double is 1023, hence we may limit - to |A| < 1023 * ln2 ~ 709. We limit to 1024 ~ 1.99 * 2^9 */ + to |A| < 1023 * ln2 ~ 709. We limit to 1024 = 2^10 */ if (f7_class_inf (a_class) - || (f7_class_nonzero (a_class) && aa->expo >= 9)) + || (f7_class_nonzero (a_class) && aa->expo >= 10)) { if (f7_class_sign (a_class)) return f7_clr (cc); diff --git a/libgcc/config/avr/libf7/libf7.h b/libgcc/config/avr/libf7/libf7.h index 8aa91c7..786e141 100644 --- a/libgcc/config/avr/libf7/libf7.h +++ b/libgcc/config/avr/libf7/libf7.h @@ -612,6 +612,7 @@ extern void f7_cos (f7_t*, const f7_t*); extern void f7_tan (f7_t*, const f7_t*); extern void f7_atan (f7_t*, const f7_t*); extern void f7_atan2 (f7_t*, const f7_t*, const f7_t*); +extern void f7_fdim (f7_t*, const f7_t*, const f7_t*); extern void f7_asin (f7_t*, const f7_t*); extern void f7_acos (f7_t*, const f7_t*); extern void f7_tanh (f7_t*, const f7_t*); diff --git a/libgfortran/io/format.c b/libgfortran/io/format.c index eef1d34..87e21a9 100644 --- a/libgfortran/io/format.c +++ b/libgfortran/io/format.c @@ -1235,9 +1235,9 @@ parse_format_list (st_parameter_dt *dtp, bool *seen_dd) default: /* Assume a missing comma with -std=legacy, GNU extension. */ - if (compile_options.warn_std == 0) - goto format_item_1; - format_error (dtp, tail, comma_missing); + if (compile_options.warn_std != 0) + fmt->error = comma_missing; + goto format_item_1; } /* Optional comma is a weird between state where we've just finished @@ -1252,7 +1252,7 @@ parse_format_list (st_parameter_dt *dtp, bool *seen_dd) case FMT_RPAREN: goto finished; - default: /* Assume that we have another format item */ + default: /* Assume that we have another format item */ fmt->saved_token = t; break; } @@ -1419,7 +1419,7 @@ parse_format (st_parameter_dt *dtp) else fmt->error = "Missing initial left parenthesis in format"; - if (format_cache_ok) + if (format_cache_ok && !fmt->error) save_parsed_format (dtp); else dtp->u.p.format_not_saved = 1; diff --git a/libgomp/ChangeLog b/libgomp/ChangeLog index 6cef1dd..c9035c3 100644 --- a/libgomp/ChangeLog +++ b/libgomp/ChangeLog @@ -1,3 +1,21 @@ +2025-05-28 Tobias Burnus <tburnus@baylibre.com> + + Backported from master: + 2025-05-28 Tobias Burnus <tburnus@baylibre.com> + + PR middle-end/118694 + * testsuite/libgomp.fortran/metadirective-1.f90: xfail when + compiling (also) for nvptx offloading as an error is then expected. + +2025-05-26 Tobias Burnus <tburnus@baylibre.com> + + Backported from master: + 2025-05-23 Tobias Burnus <tburnus@baylibre.com> + + PR middle-end/118694 + * testsuite/libgomp.c-c++-common/metadirective-1.c: xfail when + compiling (also) for nvptx offloading as an error is then expected. + 2025-05-19 Tobias Burnus <tburnus@baylibre.com> Backported from master: diff --git a/libgomp/ChangeLog.omp b/libgomp/ChangeLog.omp index 53d3bc1..5ecab6d 100644 --- a/libgomp/ChangeLog.omp +++ b/libgomp/ChangeLog.omp @@ -1,3 +1,127 @@ +2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + + * testsuite/libgomp.c++/target-flex-300.C: XFAIL. + * testsuite/libgomp.c++/target-flex-60.C: Likewise. + * testsuite/libgomp.c++/target-flex-61.C: Likewise. + * testsuite/libgomp.c++/target-flex-62.C: Likewise. + * testsuite/libgomp.c++/target-flex-81.C: Likewise. + +2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + + Backported from master: + 2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + + * testsuite/libgomp.c++/target-std__valarray-1.C: New. + * testsuite/libgomp.c++/target-std__valarray-1.output: Likewise. + +2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + + Backported from master: + 2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + + * testsuite/libgomp.c++/target-std__array-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__array-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__bitset-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__bitset-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__deque-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__deque-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__forward_list-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__forward_list-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__list-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__list-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__map-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__map-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__multimap-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__multimap-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__multiset-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__multiset-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__set-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__set-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__span-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__span-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__valarray-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__valarray-concurrent.C: Adjust. + * testsuite/libgomp.c++/target-std__vector-concurrent-usm.C: New. + * testsuite/libgomp.c++/target-std__vector-concurrent.C: Adjust. + +2025-05-30 Kwok Cheung Yeung <kcyeung@baylibre.com> + + Backported from master: + 2025-05-30 Kwok Cheung Yeung <kcyeung@baylibre.com> + Thomas Schwinge <tschwinge@baylibre.com> + + * testsuite/libgomp.c++/target-std__array-concurrent.C: New. + * testsuite/libgomp.c++/target-std__bitset-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__deque-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__flat_map-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__flat_multimap-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__flat_multiset-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__flat_set-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__forward_list-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__list-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__map-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__multimap-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__multiset-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__set-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__span-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__unordered_map-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__unordered_multimap-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__unordered_multiset-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__unordered_set-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__valarray-concurrent.C: Likewise. + * testsuite/libgomp.c++/target-std__vector-concurrent.C: Likewise. + +2025-05-30 Kwok Cheung Yeung <kcyeung@baylibre.com> + + Backported from master: + 2025-05-30 Kwok Cheung Yeung <kcyeung@baylibre.com> + + * testsuite/libgomp.c++/target-std__cmath.C: New. + * testsuite/libgomp.c++/target-std__complex.C: Likewise. + * testsuite/libgomp.c++/target-std__numbers.C: Likewise. + +2025-05-30 Waffl3x <waffl3x@baylibre.com> + + Backported from master: + 2025-05-30 Waffl3x <waffl3x@baylibre.com> + Thomas Schwinge <tschwinge@baylibre.com> + + * testsuite/libgomp.c++/target-flex-10.C: New test. + * testsuite/libgomp.c++/target-flex-100.C: New test. + * testsuite/libgomp.c++/target-flex-101.C: New test. + * testsuite/libgomp.c++/target-flex-11.C: New test. + * testsuite/libgomp.c++/target-flex-12.C: New test. + * testsuite/libgomp.c++/target-flex-2000.C: New test. + * testsuite/libgomp.c++/target-flex-2001.C: New test. + * testsuite/libgomp.c++/target-flex-2002.C: New test. + * testsuite/libgomp.c++/target-flex-2003.C: New test. + * testsuite/libgomp.c++/target-flex-30.C: New test. + * testsuite/libgomp.c++/target-flex-300.C: New test. + * testsuite/libgomp.c++/target-flex-31.C: New test. + * testsuite/libgomp.c++/target-flex-32.C: New test. + * testsuite/libgomp.c++/target-flex-33.C: New test. + * testsuite/libgomp.c++/target-flex-41.C: New test. + * testsuite/libgomp.c++/target-flex-60.C: New test. + * testsuite/libgomp.c++/target-flex-61.C: New test. + * testsuite/libgomp.c++/target-flex-62.C: New test. + * testsuite/libgomp.c++/target-flex-70.C: New test. + * testsuite/libgomp.c++/target-flex-80.C: New test. + * testsuite/libgomp.c++/target-flex-81.C: New test. + * testsuite/libgomp.c++/target-flex-90.C: New test. + * testsuite/libgomp.c++/target-flex-common.h: New test. + +2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + + Backported from master: + 2025-05-30 Thomas Schwinge <tschwinge@baylibre.com> + Richard Biener <rguenther@suse.de> + + PR middle-end/119835 + * testsuite/libgomp.oacc-c-c++-common/abi-struct-1.c: + '#pragma GCC optimize "-fno-inline"'. + * testsuite/libgomp.c-c++-common/target-abi-struct-1.c: New. + * testsuite/libgomp.c-c++-common/target-abi-struct-1-O0.c: Adjust. + 2025-05-30 Tobias Burnus <tburnus@baylibre.com> Backported from master: @@ -926,4 +1050,4 @@ * testsuite/libgomp.oacc-c-c++-common/noncontig_array-3.c: New test. * testsuite/libgomp.oacc-c-c++-common/noncontig_array-4.c: New test. * testsuite/libgomp.oacc-c-c++-common/noncontig_array-utils.h: Support - header for new tests.
\ No newline at end of file + header for new tests. diff --git a/libgomp/testsuite/libgomp.c++/target-flex-10.C b/libgomp/testsuite/libgomp.c++/target-flex-10.C new file mode 100644 index 0000000..8fa9af7 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-10.C @@ -0,0 +1,215 @@ +/* Basic container usage. */ + +#include <vector> +#include <deque> +#include <list> +#include <set> +#include <map> +#if __cplusplus >= 201103L +#include <array> +#include <forward_list> +#include <unordered_set> +#include <unordered_map> +#endif + +bool vector_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::vector<int> vector; + ok = vector.empty(); + } + return ok; +} + +bool deque_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::deque<int> deque; + ok = deque.empty(); + } + return ok; +} + +bool list_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::list<int> list; + ok = list.empty(); + } + return ok; +} + +bool map_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::map<int, int> map; + ok = map.empty(); + } + return ok; +} + +bool set_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::set<int> set; + ok = set.empty(); + } + return ok; +} + +bool multimap_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::multimap<int, int> multimap; + ok = multimap.empty(); + } + return ok; +} + +bool multiset_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::multiset<int, int> multiset; + ok = multiset.empty(); + } + return ok; +} + +#if __cplusplus >= 201103L + +bool array_test() +{ + static constexpr std::size_t array_size = 42; + bool ok; + #pragma omp target map(from: ok) + { + std::array<int, array_size> array{}; + ok = array[0] == 0 + && array[array_size - 1] == 0; + } + return ok; +} + +bool forward_list_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::forward_list<int> forward_list; + ok = forward_list.empty(); + } + return ok; +} + +bool unordered_map_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::unordered_map<int, int> unordered_map; + ok = unordered_map.empty(); + } + return ok; +} + +bool unordered_set_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + std::unordered_set<int> unordered_set; + ok = unordered_set.empty(); + } + return ok; +} + +bool unordered_multimap_test() +{ + + bool ok; + #pragma omp target map(from: ok) + { + std::unordered_multimap<int, int> unordered_multimap; + ok = unordered_multimap.empty(); + } + return ok; +} + +bool unordered_multiset_test() +{ + + bool ok; + #pragma omp target map(from: ok) + { + std::unordered_multiset<int> unordered_multiset; + ok = unordered_multiset.empty(); + } + return ok; +} + +#else +bool array_test() { return true; } +bool forward_list_test() { return true; } +bool unordered_map_test() { return true; } +bool unordered_set_test() { return true; } +bool unordered_multimap_test() { return true; } +bool unordered_multiset_test() { return true; } +#endif + +int main() +{ + const bool vec_res = vector_test(); + __builtin_printf("vector : %s\n", vec_res ? "PASS" : "FAIL"); + const bool deque_res = deque_test(); + __builtin_printf("deque : %s\n", deque_res ? "PASS" : "FAIL"); + const bool list_res = list_test(); + __builtin_printf("list : %s\n", list_res ? "PASS" : "FAIL"); + const bool map_res = map_test(); + __builtin_printf("map : %s\n", map_res ? "PASS" : "FAIL"); + const bool set_res = set_test(); + __builtin_printf("set : %s\n", set_res ? "PASS" : "FAIL"); + const bool multimap_res = multimap_test(); + __builtin_printf("multimap : %s\n", multimap_res ? "PASS" : "FAIL"); + const bool multiset_res = multiset_test(); + __builtin_printf("multiset : %s\n", multiset_res ? "PASS" : "FAIL"); + const bool array_res = array_test(); + __builtin_printf("array : %s\n", array_res ? "PASS" : "FAIL"); + const bool forward_list_res = forward_list_test(); + __builtin_printf("forward_list : %s\n", forward_list_res ? "PASS" : "FAIL"); + const bool unordered_map_res = unordered_map_test(); + __builtin_printf("unordered_map : %s\n", unordered_map_res ? "PASS" : "FAIL"); + const bool unordered_set_res = unordered_set_test(); + __builtin_printf("unordered_set : %s\n", unordered_set_res ? "PASS" : "FAIL"); + const bool unordered_multimap_res = unordered_multimap_test(); + __builtin_printf("unordered_multimap: %s\n", unordered_multimap_res ? "PASS" : "FAIL"); + const bool unordered_multiset_res = unordered_multiset_test(); + __builtin_printf("unordered_multiset: %s\n", unordered_multiset_res ? "PASS" : "FAIL"); + const bool ok = vec_res + && deque_res + && list_res + && map_res + && set_res + && multimap_res + && multiset_res + && array_res + && forward_list_res + && unordered_map_res + && unordered_set_res + && unordered_multimap_res + && unordered_multiset_res; + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-100.C b/libgomp/testsuite/libgomp.c++/target-flex-100.C new file mode 100644 index 0000000..7ab047f --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-100.C @@ -0,0 +1,210 @@ +/* Container adaptors in target region. + Does not test comparison operators other than equality to allow these tests + to be generalized to arbitrary input data. */ + +#include <algorithm> +#include <cstdio> +#include <deque> +#include <queue> +#include <stack> +#include <vector> + +#include "target-flex-common.h" + +template<typename T, std::size_t Size> +bool test_stack(T (&arr)[Size]) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + bool inner_ok = true; + const std::size_t half_size = Size / 2; + const T first_element = arr[0]; + const T middle_element = arr[half_size - 1]; + const T last_element = arr[Size - 1]; + typedef std::stack<T, std::vector<T> > stack_type; + stack_type stack; + VERIFY (stack.empty()); + VERIFY (stack.size() == 0); + { + /* Do half with push. */ + std::size_t idx = 0; + for (; idx < half_size; ++idx) + { + stack.push(arr[idx]); + VERIFY (stack.top() == arr[idx]); + } + VERIFY (stack.size() == half_size); + VERIFY (static_cast<const stack_type&>(stack).size() == half_size); + for (; idx < Size; ++idx) + { + #if __cplusplus >= 201103L + /* Do the rest with emplace if C++11 or higher. */ + stack.emplace(arr[idx]); + #else + /* Otherwise just use push again. */ + stack.push(arr[idx]); + #endif + VERIFY (stack.top() == arr[idx]); + } + VERIFY (stack.size() == Size); + VERIFY (static_cast<const stack_type&>(stack).size() == Size); + + const stack_type stack_orig = stack_type(std::vector<T>(arr, arr + Size)); + VERIFY (stack == stack_orig); + /* References are contained in their own scope so we don't accidently + add tests referencing them after they have been invalidated. */ + { + const T& const_top = static_cast<const stack_type&>(stack).top(); + VERIFY (const_top == last_element); + T& mutable_top = stack.top(); + mutable_top = first_element; + VERIFY (const_top == first_element); + } + /* Will only compare inequal if the first and last elements are different. */ + VERIFY (first_element != last_element || stack != stack_orig); + for (std::size_t count = Size - half_size; count != 0; --count) + stack.pop(); + VERIFY (stack.top() == middle_element); + const stack_type stack_half_orig = stack_type(std::vector<T>(arr, arr + half_size)); + VERIFY (stack == stack_half_orig); + } + end: + ok = inner_ok; + } + return ok; +} + +template<typename T, std::size_t Size> +bool test_queue(T (&arr)[Size]) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + bool inner_ok = true; + const std::size_t half_size = Size / 2; + const T first_element = arr[0]; + const T last_element = arr[Size - 1]; + typedef std::queue<T, std::deque<T> > queue_type; + queue_type queue; + VERIFY (queue.empty()); + VERIFY (queue.size() == 0); + { + /* Do half with push. */ + std::size_t idx = 0; + for (; idx < half_size; ++idx) + { + queue.push(arr[idx]); + VERIFY (queue.back() == arr[idx]); + VERIFY (queue.front() == first_element); + } + VERIFY (queue.size() == half_size); + VERIFY (static_cast<const queue_type&>(queue).size() == half_size); + for (; idx < Size; ++idx) + { + #if __cplusplus >= 201103L + /* Do the rest with emplace if C++11 or higher. */ + queue.emplace(arr[idx]); + #else + /* Otherwise just use push again. */ + queue.push(arr[idx]); + #endif + VERIFY (queue.back() == arr[idx]); + } + VERIFY (queue.size() == Size); + VERIFY (static_cast<const queue_type&>(queue).size() == Size); + + const queue_type queue_orig = queue_type(std::deque<T>(arr, arr + Size)); + VERIFY (queue == queue_orig); + + /* References are contained in their own scope so we don't accidently + add tests referencing them after they have been invalidated. */ + { + const T& const_front = static_cast<const queue_type&>(queue).front(); + VERIFY (const_front == first_element); + T& mutable_front = queue.front(); + + const T& const_back = static_cast<const queue_type&>(queue).back(); + VERIFY (const_back == last_element); + T& mutable_back = queue.back(); + { + using std::swap; + swap(mutable_front, mutable_back); + } + VERIFY (const_front == last_element); + VERIFY (const_back == first_element); + /* Will only compare inequal if the first and last elements are different. */ + VERIFY (first_element != last_element || queue != queue_orig); + /* Return the last element to normal for the next comparison. */ + mutable_back = last_element; + } + + const T middle_element = arr[half_size]; + for (std::size_t count = Size - half_size; count != 0; --count) + queue.pop(); + VERIFY (queue.front() == middle_element); + const queue_type queue_upper_half = queue_type(std::deque<T>(arr + half_size, arr + Size)); + VERIFY (queue == queue_upper_half); + } + end: + ok = inner_ok; + } + return ok; +} + +template<typename T, std::size_t Size> +bool test_priority_queue(T (&arr)[Size], const T min_value, const T max_value) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + bool inner_ok = true; + typedef std::priority_queue<T, std::vector<T> > priority_queue_type; + { + priority_queue_type pqueue; + VERIFY (pqueue.empty()); + VERIFY (pqueue.size() == 0); + } + { + priority_queue_type pqueue(arr, arr + Size); + VERIFY (!pqueue.empty()); + VERIFY (pqueue.size() == Size); + VERIFY (static_cast<const priority_queue_type&>(pqueue).size() == Size); + + const T old_max = pqueue.top(); + + #if __cplusplus >= 201103L + pqueue.emplace(max_value); + #else + pqueue.push(max_value); + #endif + VERIFY (pqueue.top() == max_value); + pqueue.pop(); + VERIFY (pqueue.top() == old_max); + pqueue.push(min_value); + VERIFY (pqueue.top() == old_max); + pqueue.push(max_value); + VERIFY (pqueue.top() == max_value); + pqueue.pop(); + VERIFY (pqueue.top() == old_max); + VERIFY (pqueue.size() == Size + 1); + + for (std::size_t count = Size; count != 0; --count) + pqueue.pop(); + VERIFY (pqueue.size() == 1); + VERIFY (pqueue.top() == min_value); + } + end: + ok = inner_ok; + } + return ok; +} + +int main() +{ + int arr[10] = {0,1,2,3,4,5,6,7,8,9}; + + return test_stack(arr) + && test_queue(arr) + && test_priority_queue(arr, 0, 1000) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-101.C b/libgomp/testsuite/libgomp.c++/target-flex-101.C new file mode 100644 index 0000000..be9037e --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-101.C @@ -0,0 +1,136 @@ +/* { dg-additional-options -std=c++23 } */ + +/* C++23 container adaptors in target region. + Severely needs additional tests. */ + +#include <cstdio> +#include <utility> +#include <version> + +#if __cpp_lib_flat_map >= 202207L +#define ENABLE_FLAT_MAP 1 +#endif +#if __cpp_lib_flat_set >= 202207L +#define ENABLE_FLAT_SET 1 +#endif + +#ifdef ENABLE_FLAT_MAP +#include <flat_map> +#endif +#ifdef ENABLE_FLAT_SET +#include <flat_set> +#endif + +#include "target-flex-common.h" + +#ifdef ENABLE_FLAT_MAP +template<typename K, typename V, typename std::size_t Size> +bool test_flat_map(std::pair<K, V> (&arr)[Size]) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + bool inner_ok = true; + { + using flat_map_type = std::flat_map<K, V>; + flat_map_type map = {arr, arr + Size}; + + VERIFY (!map.empty()); + for (const auto& element : arr) + VERIFY (map.contains(element.first)); + } + end: + ok = inner_ok; + } + return ok; +} + +template<typename K, typename V, typename std::size_t Size> +bool test_flat_multimap(std::pair<K, V> (&arr)[Size]) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + bool inner_ok = true; + { + using flat_map_type = std::flat_map<K, V>; + flat_map_type map = {arr, arr + Size}; + + VERIFY (!map.empty()); + for (const auto& element : arr) + VERIFY (map.contains(element.first)); + } + end: + ok = inner_ok; + } + return ok; +} +#else +template<typename K, typename V, typename std::size_t Size> +bool test_flat_map(std::pair<K, V> (&arr)[Size]) { return true; } + +template<typename K, typename V, typename std::size_t Size> +bool test_flat_multimap(std::pair<K, V> (&arr)[Size]) { return true; } +#endif + +#ifdef ENABLE_FLAT_SET +template<typename T, typename std::size_t Size> +bool test_flat_set(T (&arr)[Size]) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + bool inner_ok = true; + { + using flat_set_type = std::flat_set<T>; + flat_set_type set = {arr, arr + Size}; + + VERIFY (!set.empty()); + for (const auto& element : arr) + VERIFY (set.contains(element)); + } + end: + ok = inner_ok; + } + return ok; +} + +template<typename T, typename std::size_t Size> +bool test_flat_multiset(T (&arr)[Size]) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + bool inner_ok = true; + { + using flat_multiset_type = std::flat_multiset<T>; + flat_multiset_type multiset = {arr, arr + Size}; + + VERIFY (!multiset.empty()); + for (const auto& element : arr) + VERIFY (multiset.contains(element)); + } + end: + ok = inner_ok; + } + return ok; +} +#else +template<typename T, typename std::size_t Size> +bool test_flat_set(T (&arr)[Size]) { return true; } + +template<typename T, typename std::size_t Size> +bool test_flat_multiset(T (&arr)[Size]) { return true; } +#endif + +int main() +{ + int arr[10] = {0,1,2,3,4,5,6,7,8,9}; + std::pair<int, int> pairs[10] = {{ 1, 2}, { 2, 4}, { 3, 6}, { 4, 8}, { 5, 10}, + { 6, 12}, { 7, 14}, { 8, 16}, { 9, 18}, {10, 20}}; + + return test_flat_set(arr) + && test_flat_multiset(arr) + && test_flat_map(pairs) + && test_flat_multimap(pairs) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-11.C b/libgomp/testsuite/libgomp.c++/target-flex-11.C new file mode 100644 index 0000000..6d55129 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-11.C @@ -0,0 +1,444 @@ +/* Check constructors/destructors are called in containers. */ + +#include <vector> +#include <deque> +#include <list> +#include <set> +#include <map> +#include <utility> +#if __cplusplus >= 201103L +#include <array> +#include <forward_list> +#include <unordered_set> +#include <unordered_map> +#endif + +#include "target-flex-common.h" + +struct indirect_counter +{ + typedef int counter_value_type; + counter_value_type *_count_ptr; + + indirect_counter(counter_value_type *count_ptr) BL_NOEXCEPT : _count_ptr(count_ptr) { + ++(*_count_ptr); + } + indirect_counter(const indirect_counter& other) BL_NOEXCEPT : _count_ptr(other._count_ptr) { + ++(*_count_ptr); + } + /* Don't declare a move constructor, we want to copy no matter what. */ + ~indirect_counter() { + --(*_count_ptr); + } +}; + +bool operator==(indirect_counter const& lhs, indirect_counter const& rhs) BL_NOEXCEPT + { return lhs._count_ptr == rhs._count_ptr; } +bool operator<(indirect_counter const& lhs, indirect_counter const& rhs) BL_NOEXCEPT + { return lhs._count_ptr < rhs._count_ptr; } + +#if __cplusplus >= 201103L +template<> +struct std::hash<indirect_counter> +{ + std::size_t operator()(const indirect_counter& ic) const noexcept + { return std::hash<indirect_counter::counter_value_type *>{}(ic._count_ptr); } +}; +#endif + +/* Not a container, just a sanity check really. */ +bool automatic_lifetime_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + indirect_counter c = indirect_counter(&counter); + indirect_counter(static_cast<int*>(&counter)); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool vector_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::vector<indirect_counter> vec(42, indirect_counter(&counter)); + VERIFY (counter == 42); + vec.resize(32, indirect_counter(&counter)); + VERIFY (counter == 32); + vec.push_back(indirect_counter(&counter)); + VERIFY (counter == 33); + vec.pop_back(); + VERIFY (counter == 32); + vec.pop_back(); + VERIFY (counter == 31); + vec.resize(100, indirect_counter(&counter)); + VERIFY (counter == 100); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool deque_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::deque<indirect_counter> vec(42, indirect_counter(&counter)); + VERIFY (counter == 42); + vec.resize(32, indirect_counter(&counter)); + VERIFY (counter == 32); + vec.push_back(indirect_counter(&counter)); + VERIFY (counter == 33); + vec.pop_back(); + VERIFY (counter == 32); + vec.pop_back(); + VERIFY (counter == 31); + vec.resize(100, indirect_counter(&counter)); + VERIFY (counter == 100); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool list_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::list<indirect_counter> list(42, indirect_counter(&counter)); + VERIFY (counter == 42); + list.resize(32, indirect_counter(&counter)); + VERIFY (counter == 32); + list.push_back(indirect_counter(&counter)); + VERIFY (counter == 33); + list.pop_back(); + VERIFY (counter == 32); + list.pop_back(); + VERIFY (counter == 31); + list.resize(100, indirect_counter(&counter)); + VERIFY (counter == 100); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool map_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::map<int, indirect_counter> map; + map.insert(std::make_pair(1, indirect_counter(&counter))); + VERIFY (counter == 1); + map.insert(std::make_pair(1, indirect_counter(&counter))); + VERIFY (counter == 1); + map.insert(std::make_pair(2, indirect_counter(&counter))); + VERIFY (counter == 2); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool set_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter0 = 0; + int counter1 = 0; + { + std::set<indirect_counter> set; + set.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 1); + set.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 1); + set.insert(indirect_counter(&counter1)); + VERIFY (counter0 == 1 && counter1 == 1); + } + VERIFY (counter0 == 0 && counter1 == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool multimap_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::multimap<int, indirect_counter> multimap; + multimap.insert(std::make_pair(1, indirect_counter(&counter))); + VERIFY (counter == 1); + multimap.insert(std::make_pair(1, indirect_counter(&counter))); + VERIFY (counter == 2); + multimap.insert(std::make_pair(2, indirect_counter(&counter))); + VERIFY (counter == 3); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool multiset_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter0 = 0; + int counter1 = 0; + { + std::multiset<indirect_counter> multiset; + multiset.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 1); + multiset.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 2); + multiset.insert(indirect_counter(&counter1)); + VERIFY (counter0 == 2 && counter1 == 1); + } + VERIFY (counter0 == 0 && counter1 == 0); + end: + ok = inner_ok; + } + return ok; +} + +#if __cplusplus >= 201103L + +bool array_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + indirect_counter ic(&counter); + std::array<indirect_counter, 10> array{ic, ic, ic, ic, ic, + ic, ic, ic, ic, ic}; + VERIFY (counter == 11); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool forward_list_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::forward_list<indirect_counter> forward_list(42, indirect_counter(&counter)); + VERIFY (counter == 42); + forward_list.resize(32, indirect_counter(&counter)); + VERIFY (counter == 32); + forward_list.push_front(indirect_counter(&counter)); + VERIFY (counter == 33); + forward_list.pop_front(); + VERIFY (counter == 32); + forward_list.pop_front(); + VERIFY (counter == 31); + forward_list.resize(100, indirect_counter(&counter)); + VERIFY (counter == 100); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool unordered_map_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::unordered_map<int, indirect_counter> unordered_map; + unordered_map.insert({1, indirect_counter(&counter)}); + VERIFY (counter == 1); + unordered_map.insert({1, indirect_counter(&counter)}); + VERIFY (counter == 1); + unordered_map.insert({2, indirect_counter(&counter)}); + VERIFY (counter == 2); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool unordered_set_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter0 = 0; + int counter1 = 0; + { + std::unordered_set<indirect_counter> unordered_set; + unordered_set.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 1); + unordered_set.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 1); + unordered_set.insert(indirect_counter(&counter1)); + VERIFY (counter0 == 1 && counter1 == 1); + } + VERIFY (counter0 == 0 && counter1 == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool unordered_multimap_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter = 0; + { + std::unordered_multimap<int, indirect_counter> unordered_multimap; + unordered_multimap.insert({1, indirect_counter(&counter)}); + VERIFY (counter == 1); + unordered_multimap.insert({1, indirect_counter(&counter)}); + VERIFY (counter == 2); + unordered_multimap.insert({2, indirect_counter(&counter)}); + VERIFY (counter == 3); + } + VERIFY (counter == 0); + end: + ok = inner_ok; + } + return ok; +} + +bool unordered_multiset_test() +{ + bool ok; + #pragma omp target map(from: ok) + { + bool inner_ok = true; + int counter0 = 0; + int counter1 = 0; + { + std::unordered_multiset<indirect_counter> unordered_multiset; + unordered_multiset.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 1); + unordered_multiset.insert(indirect_counter(&counter0)); + VERIFY (counter0 == 2); + unordered_multiset.insert(indirect_counter(&counter1)); + VERIFY (counter0 == 2 && counter1 == 1); + } + VERIFY (counter0 == 0 && counter1 == 0); + end: + ok = inner_ok; + } + return ok; +} + +#else +bool array_test() { return true; } +bool forward_list_test() { return true; } +bool unordered_map_test() { return true; } +bool unordered_set_test() { return true; } +bool unordered_multimap_test() { return true; } +bool unordered_multiset_test() { return true; } +#endif + +int main() +{ + const bool auto_res = automatic_lifetime_test(); + const bool vec_res = vector_test(); + const bool deque_res = deque_test(); + const bool list_res = list_test(); + const bool map_res = map_test(); + const bool set_res = set_test(); + const bool multimap_res = multimap_test(); + const bool multiset_res = multiset_test(); + const bool array_res = array_test(); + const bool forward_list_res = forward_list_test(); + const bool unordered_map_res = unordered_map_test(); + const bool unordered_set_res = unordered_set_test(); + const bool unordered_multimap_res = unordered_multimap_test(); + const bool unordered_multiset_res = unordered_multiset_test(); + std::printf("sanity check : %s\n", auto_res ? "PASS" : "FAIL"); + std::printf("vector : %s\n", vec_res ? "PASS" : "FAIL"); + std::printf("deque : %s\n", deque_res ? "PASS" : "FAIL"); + std::printf("list : %s\n", list_res ? "PASS" : "FAIL"); + std::printf("map : %s\n", map_res ? "PASS" : "FAIL"); + std::printf("set : %s\n", set_res ? "PASS" : "FAIL"); + std::printf("multimap : %s\n", multimap_res ? "PASS" : "FAIL"); + std::printf("multiset : %s\n", multiset_res ? "PASS" : "FAIL"); + std::printf("array : %s\n", array_res ? "PASS" : "FAIL"); + std::printf("forward_list : %s\n", forward_list_res ? "PASS" : "FAIL"); + std::printf("unordered_map : %s\n", unordered_map_res ? "PASS" : "FAIL"); + std::printf("unordered_set : %s\n", unordered_set_res ? "PASS" : "FAIL"); + std::printf("unordered_multimap: %s\n", unordered_multimap_res ? "PASS" : "FAIL"); + std::printf("unordered_multiset: %s\n", unordered_multiset_res ? "PASS" : "FAIL"); + const bool ok = auto_res + && vec_res + && deque_res + && list_res + && map_res + && set_res + && multimap_res + && multiset_res + && array_res + && forward_list_res + && unordered_map_res + && unordered_set_res + && unordered_multimap_res + && unordered_multiset_res; + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-12.C b/libgomp/testsuite/libgomp.c++/target-flex-12.C new file mode 100644 index 0000000..024fb73 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-12.C @@ -0,0 +1,736 @@ +/* Populated with mapped data, validate, mutate, validate again. + The cases using sets do not mutate. + Note: Some of the code in here really sucks due to being made to be + compatible with c++98. */ + +#include <vector> +#include <deque> +#include <list> +#include <set> +#include <map> +#if __cplusplus >= 201103L +#include <array> +#include <forward_list> +#include <unordered_set> +#include <unordered_map> +#endif + +#include <limits> +#include <iterator> + +#include "target-flex-common.h" + +template<bool B, class T = void> +struct enable_if {}; + +template<class T> +struct enable_if<true, T> { typedef T type; }; + +struct identity_func +{ +#if __cplusplus < 201103L + template<typename T> + T& operator()(T& arg) const BL_NOEXCEPT { return arg; } + template<typename T> + T const& operator()(T const& arg) const BL_NOEXCEPT { return arg; } +#else + template<typename T> + constexpr T&& operator()(T&& arg) const BL_NOEXCEPT { return std::forward<T>(arg); } +#endif +}; + +/* Applies projection to the second iterator. */ +template<typename It0, typename It1, typename Proj> +bool validate_sequential_elements(const It0 begin0, const It0 end0, + const It1 begin1, const It1 end1, + Proj proj) BL_NOEXCEPT +{ + It0 it0 = begin0; + It1 it1 = begin1; + for (; it0 != end0; ++it0, ++it1) + { + /* Sizes mismatch, don't bother aborting though just fail the test. */ + if (it1 == end1) + return false; + if (*it0 != proj(*it1)) + return false; + } + /* Sizes mismatch, do as above. */ + if (it1 != end1) + return false; + return true; +} + +template<typename It0, typename It1> +bool validate_sequential_elements(const It0 begin0, const It0 end0, + const It1 begin1, const It1 end1) BL_NOEXCEPT +{ + return validate_sequential_elements(begin0, end0, begin1, end1, identity_func()); +} + +/* Inefficient, but simple. */ +template<typename It, typename OutIt> +void simple_copy(const It begin, const It end, OutIt out) BL_NOEXCEPT +{ + for (It it = begin; it != end; ++it, ++out) + *out = *it; +} + +template<typename It, typename MutateFn> +void simple_mutate(const It begin, const It end, MutateFn mut_fn) BL_NOEXCEPT +{ + for (It it = begin; it != end; ++it) + *it = mut_fn(*it); +} + +template<typename MutationFunc, typename T, std::size_t Size> +bool vector_test(const T (&arr)[Size]) +{ + bool ok; + T out_arr[Size]; + T out_mut_arr[Size]; + #pragma omp target map(from: ok, out_arr[:Size], out_mut_arr[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::vector<T> vector(arr, arr + Size); + VERIFY (validate_sequential_elements(vector.begin(), vector.end(), + arr, arr + Size)); + simple_copy(vector.begin(), vector.end(), out_arr); + simple_mutate(vector.begin(), vector.end(), MutationFunc()); + VERIFY (validate_sequential_elements(vector.begin(), vector.end(), + arr, arr + Size, MutationFunc())); + simple_copy(vector.begin(), vector.end(), out_mut_arr); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_sequential_elements(out_arr, out_arr + Size, + arr, arr + Size)); + VERIFY_NON_TARGET (validate_sequential_elements(out_mut_arr, out_mut_arr + Size, + arr, arr + Size, MutationFunc())); + return true; +} + +template<typename MutationFunc, typename T, std::size_t Size> +bool deque_test(const T (&arr)[Size]) +{ + bool ok; + T out_arr[Size]; + T out_mut_arr[Size]; + #pragma omp target map(from: ok, out_arr[:Size], out_mut_arr[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::deque<T> deque(arr, arr + Size); + VERIFY (validate_sequential_elements(deque.begin(), deque.end(), + arr, arr + Size)); + simple_copy(deque.begin(), deque.end(), out_arr); + simple_mutate(deque.begin(), deque.end(), MutationFunc()); + VERIFY (validate_sequential_elements(deque.begin(), deque.end(), + arr, arr + Size, MutationFunc())); + simple_copy(deque.begin(), deque.end(), out_mut_arr); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_sequential_elements(out_arr, out_arr + Size, + arr, arr + Size)); + VERIFY_NON_TARGET (validate_sequential_elements(out_mut_arr, out_mut_arr + Size, + arr, arr + Size, MutationFunc())); + return true; +} + +template<typename MutationFunc, typename T, std::size_t Size> +bool list_test(const T (&arr)[Size]) +{ + bool ok; + T out_arr[Size]; + T out_mut_arr[Size]; + #pragma omp target map(from: ok, out_arr[:Size], out_mut_arr[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::list<T> list(arr, arr + Size); + VERIFY (validate_sequential_elements(list.begin(), list.end(), + arr, arr + Size)); + simple_copy(list.begin(), list.end(), out_arr); + simple_mutate(list.begin(), list.end(), MutationFunc()); + VERIFY (validate_sequential_elements(list.begin(), list.end(), + arr, arr + Size, MutationFunc())); + simple_copy(list.begin(), list.end(), out_mut_arr); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_sequential_elements(out_arr, out_arr + Size, + arr, arr + Size)); + VERIFY_NON_TARGET (validate_sequential_elements(out_mut_arr, out_mut_arr + Size, + arr, arr + Size, MutationFunc())); + return true; +} + +template<typename T> +const T& get_key(const T& arg) BL_NOEXCEPT + { return arg; } +template<typename K, typename V> +const K& get_key(const std::pair<K, V>& pair) BL_NOEXCEPT + { return pair.first; } +template<typename T> +const T& get_value(const T& arg) BL_NOEXCEPT + { return arg; } +template<typename K, typename V> +const K& get_value(const std::pair<K, V>& pair) BL_NOEXCEPT + { return pair.second; } + +template<typename T> +struct key_type { typedef T type; }; +template<typename K, typename V> +struct key_type<std::pair<K, V> > { typedef K type; }; + +template<typename Proj, typename Container, typename It> +bool validate_associative(const Container& container, + const It compare_begin, + const It compare_end, + Proj proj) BL_NOEXCEPT +{ + const typename Container::const_iterator elem_end = container.end(); + for (It compare_it = compare_begin; compare_it != compare_end; ++compare_it) + { + const typename Container::const_iterator elem_it = container.find(get_key(*compare_it)); + VERIFY_NON_TARGET (elem_it != elem_end); + VERIFY_NON_TARGET (proj(get_value(*compare_it)) == get_value(*elem_it)); + } + return true; +} + +template<typename Container, typename It> +bool validate_associative(const Container& container, + const It compare_begin, + const It compare_end) BL_NOEXCEPT +{ + return validate_associative(container, compare_begin, compare_end, identity_func()); +} + +template<typename It, typename MutateFn> +void simple_mutate_map(const It begin, const It end, MutateFn mut_fn) BL_NOEXCEPT +{ + for (It it = begin; it != end; ++it) + it->second = mut_fn(it->second); +} + +template<typename It, typename OutIter> +void simple_copy_unique(const It begin, const It end, OutIter out) BL_NOEXCEPT +{ + /* In case anyone reads this, I want it to be known that I hate c++98. */ + typedef typename key_type<typename std::iterator_traits<It>::value_type>::type key_t; + std::set<key_t> already_seen; + for (It it = begin; it != end; ++it, ++out) + { + key_t key = get_key(*it); + if (already_seen.find(key) != already_seen.end()) + continue; + already_seen.insert(key); + *out = *it; + } +} + +template<typename MutationFunc, typename K, typename V, std::size_t Size> +bool map_test(const std::pair<K, V> (&arr)[Size]) +{ + std::map<K, V> reference_map(arr, arr + Size); + bool ok; + /* Both sizes should be the same. */ + std::pair<K, V> out_pairs[Size]; + std::size_t out_size; + std::pair<K, V> out_pairs_mut[Size]; + std::size_t out_size_mut; + #pragma omp target map(from: ok, out_pairs[:Size], out_size, \ + out_pairs_mut[:Size], out_size_mut) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::vector<std::pair<K, V> > unique_elems; + simple_copy_unique(arr, arr + Size, + std::back_insert_iterator<std::vector<std::pair<K, V> > >(unique_elems)); + + std::map<K, V> map(arr, arr + Size); + VERIFY (validate_associative(map, unique_elems.begin(), unique_elems.end())); + simple_copy(map.begin(), map.end(), out_pairs); + out_size = map.size(); + simple_mutate_map(map.begin(), map.end(), MutationFunc()); + VERIFY (validate_associative(map, unique_elems.begin(), unique_elems.end(), + MutationFunc())); + simple_copy(map.begin(), map.end(), out_pairs_mut); + out_size_mut = map.size(); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (out_size == out_size_mut); + VERIFY_NON_TARGET (validate_associative(reference_map, + out_pairs, out_pairs + out_size)); + simple_mutate_map(reference_map.begin(), reference_map.end(), MutationFunc()); + VERIFY_NON_TARGET (validate_associative(reference_map, + out_pairs_mut, out_pairs_mut + out_size_mut)); + return true; +} + +template<typename T, std::size_t Size> +bool set_test(const T (&arr)[Size]) +{ + std::set<T> reference_set(arr, arr + Size); + bool ok; + /* Both sizes should be the same. */ + T out_arr[Size]; + std::size_t out_size; + #pragma omp target map(from: ok, out_arr[:Size], out_size) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::vector<T> unique_elems; + simple_copy_unique(arr, arr + Size, + std::back_insert_iterator<std::vector<T> >(unique_elems)); + + std::set<T> set(arr, arr + Size); + VERIFY (validate_associative(set, unique_elems.begin(), unique_elems.end())); + simple_copy(set.begin(), set.end(), out_arr); + out_size = set.size(); + /* Sets can't be mutated, we could create another set with mutated + but it gets a little annoying and probably isn't an interesting test. */ + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_associative(reference_set, + out_arr, out_arr + out_size)); + return true; +} + +template<typename Proj, typename Container, typename It> +bool validate_multi_associative(const Container& container, + const It compare_begin, + const It compare_end, + Proj proj) BL_NOEXCEPT +{ + /* Once again, for the poor soul reviewing these, I hate c++98. */ + typedef typename key_type<typename std::iterator_traits<It>::value_type>::type key_t; + typedef std::map<key_t, std::size_t> counter_map; + counter_map key_count_map; + for (It it = compare_begin; it != compare_end; ++it) + { + const key_t& key = get_key(*it); + typename counter_map::iterator counter_it + = key_count_map.find(key); + if (counter_it != key_count_map.end()) + ++counter_it->second; + else + key_count_map.insert(std::pair<const key_t, std::size_t>(key, std::size_t(1))); + } + const typename Container::const_iterator elem_end = container.end(); + for (It compare_it = compare_begin; compare_it != compare_end; ++compare_it) + { + const key_t& key = get_key(*compare_it); + typename counter_map::iterator count_it = key_count_map.find(key); + std::size_t key_count = count_it != key_count_map.end() ? count_it->second + : std::size_t(0); + VERIFY_NON_TARGET (key_count > std::size_t(0) && "this will never happen"); + /* This gets tested multiple times but that should be fine. */ + VERIFY_NON_TARGET (key_count == container.count(key)); + typename Container::const_iterator elem_it = container.find(key); + /* This will never happen if the previous case passed. */ + VERIFY_NON_TARGET (elem_it != elem_end); + bool found_element = false; + for (; elem_it != elem_end; ++elem_it) + if (proj(get_value(*compare_it)) == get_value(*elem_it)) + { + found_element = true; + break; + } + VERIFY_NON_TARGET (found_element); + } + return true; +} + +template<typename Container, typename It> +bool validate_multi_associative(const Container& container, + const It compare_begin, + const It compare_end) BL_NOEXCEPT +{ + return validate_multi_associative(container, compare_begin, compare_end, identity_func()); +} + +template<typename MutationFunc, typename K, typename V, std::size_t Size> +bool multimap_test(const std::pair<K, V> (&arr)[Size]) +{ + std::multimap<K, V> reference_multimap(arr, arr + Size); + bool ok; + std::pair<K, V> out_pairs[Size]; + std::pair<K, V> out_pairs_mut[Size]; + #pragma omp target map(from: ok, out_pairs[:Size], out_pairs_mut[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::multimap<K, V> multimap(arr, arr + Size); + VERIFY (validate_multi_associative(multimap, arr, arr + Size)); + simple_copy(multimap.begin(), multimap.end(), out_pairs); + simple_mutate_map(multimap.begin(), multimap.end(), MutationFunc()); + VERIFY (validate_multi_associative(multimap, arr, arr + Size, MutationFunc())); + simple_copy(multimap.begin(), multimap.end(), out_pairs_mut); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_multi_associative(reference_multimap, + out_pairs, out_pairs + Size)); + simple_mutate_map(reference_multimap.begin(), reference_multimap.end(), MutationFunc()); + VERIFY_NON_TARGET (validate_multi_associative(reference_multimap, + out_pairs_mut, out_pairs_mut + Size)); + return true; +} + +template<typename T, std::size_t Size> +bool multiset_test(const T (&arr)[Size]) +{ + std::multiset<T> reference_multiset(arr, arr + Size); + bool ok; + T out_arr[Size]; + #pragma omp target map(from: ok, out_arr[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::multiset<T> set(arr, arr + Size); + VERIFY (validate_multi_associative(set, arr, arr + Size)); + simple_copy(set.begin(), set.end(), out_arr); + /* Sets can't be mutated, we could create another set with mutated + but it gets a little annoying and probably isn't an interesting test. */ + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_multi_associative(reference_multiset, + out_arr, out_arr + Size)); + return true; +} + +#if __cplusplus >= 201103L + +template<typename MutationFunc, typename T, std::size_t Size> +bool array_test(const T (&arr)[Size]) +{ + bool ok; + T out_arr[Size]; + T out_mut_arr[Size]; + #pragma omp target map(from: ok, out_arr[:Size], out_mut_arr[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::array<T, Size> std_array{}; + /* Special case for std::array since it can't be initialized + with iterators. */ + { + T zero_val = T{}; + for (auto it = std_array.begin(); it != std_array.end(); ++it) + VERIFY (*it == zero_val); + } + simple_copy(arr, arr + Size, std_array.begin()); + VERIFY (validate_sequential_elements(std_array.begin(), std_array.end(), + arr, arr + Size)); + simple_copy(std_array.begin(), std_array.end(), out_arr); + simple_mutate(std_array.begin(), std_array.end(), MutationFunc()); + VERIFY (validate_sequential_elements(std_array.begin(), std_array.end(), + arr, arr + Size, MutationFunc())); + simple_copy(std_array.begin(), std_array.end(), out_mut_arr); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_sequential_elements(out_arr, out_arr + Size, + arr, arr + Size)); + VERIFY_NON_TARGET (validate_sequential_elements(out_mut_arr, out_mut_arr + Size, + arr, arr + Size, MutationFunc())); + return true; +} + +template<typename MutationFunc, typename T, std::size_t Size> +bool forward_list_test(const T (&arr)[Size]) +{ + bool ok; + T out_arr[Size]; + T out_mut_arr[Size]; + #pragma omp target map(from: ok, out_arr[:Size], out_mut_arr[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::forward_list<T> fwd_list(arr, arr + Size); + VERIFY (validate_sequential_elements(fwd_list.begin(), fwd_list.end(), + arr, arr + Size)); + simple_copy(fwd_list.begin(), fwd_list.end(), out_arr); + simple_mutate(fwd_list.begin(), fwd_list.end(), MutationFunc()); + VERIFY (validate_sequential_elements(fwd_list.begin(), fwd_list.end(), + arr, arr + Size, MutationFunc())); + simple_copy(fwd_list.begin(), fwd_list.end(), out_mut_arr); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_sequential_elements(out_arr, out_arr + Size, + arr, arr + Size)); + VERIFY_NON_TARGET (validate_sequential_elements(out_mut_arr, out_mut_arr + Size, + arr, arr + Size, MutationFunc())); + return true; +} + +template<typename MutationFunc, typename K, typename V, std::size_t Size> +bool unordered_map_test(const std::pair<K, V> (&arr)[Size]) +{ + std::unordered_map<K, V> reference_map(arr, arr + Size); + bool ok; + /* Both sizes should be the same. */ + std::pair<K, V> out_pairs[Size]; + std::size_t out_size; + std::pair<K, V> out_pairs_mut[Size]; + std::size_t out_size_mut; + #pragma omp target map(from: ok, out_pairs[:Size], out_size, \ + out_pairs_mut[:Size], out_size_mut) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::vector<std::pair<K, V> > unique_elems; + simple_copy_unique(arr, arr + Size, + std::back_insert_iterator<std::vector<std::pair<K, V> > >(unique_elems)); + + std::unordered_map<K, V> map(arr, arr + Size); + VERIFY (validate_associative(map, unique_elems.begin(), unique_elems.end())); + simple_copy(map.begin(), map.end(), out_pairs); + out_size = map.size(); + simple_mutate_map(map.begin(), map.end(), MutationFunc()); + VERIFY (validate_associative(map, unique_elems.begin(), unique_elems.end(), + MutationFunc())); + simple_copy(map.begin(), map.end(), out_pairs_mut); + out_size_mut = map.size(); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (out_size == out_size_mut); + VERIFY_NON_TARGET (validate_associative(reference_map, + out_pairs, out_pairs + out_size)); + simple_mutate_map(reference_map.begin(), reference_map.end(), MutationFunc()); + VERIFY_NON_TARGET (validate_associative(reference_map, + out_pairs_mut, out_pairs_mut + out_size_mut)); + return true; +} + +template<typename T, std::size_t Size> +bool unordered_set_test(const T (&arr)[Size]) +{ + std::unordered_set<T> reference_set(arr, arr + Size); + bool ok; + /* Both sizes should be the same. */ + T out_arr[Size]; + std::size_t out_size; + #pragma omp target map(from: ok, out_arr[:Size], out_size) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::vector<T> unique_elems; + simple_copy_unique(arr, arr + Size, + std::back_insert_iterator<std::vector<T> >(unique_elems)); + + std::unordered_set<T> set(arr, arr + Size); + VERIFY (validate_associative(set, unique_elems.begin(), unique_elems.end())); + simple_copy(set.begin(), set.end(), out_arr); + out_size = set.size(); + /* Sets can't be mutated, we could create another set with mutated + but it gets a little annoying and probably isn't an interesting test. */ + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_associative(reference_set, + out_arr, out_arr + out_size)); + return true; +} + +template<typename MutationFunc, typename K, typename V, std::size_t Size> +bool unordered_multimap_test(const std::pair<K, V> (&arr)[Size]) +{ + std::unordered_multimap<K, V> reference_multimap(arr, arr + Size); + bool ok; + std::pair<K, V> out_pairs[Size]; + std::pair<K, V> out_pairs_mut[Size]; + #pragma omp target map(from: ok, out_pairs[:Size], out_pairs_mut[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::unordered_multimap<K, V> multimap(arr, arr + Size); + VERIFY (validate_multi_associative(multimap, arr, arr + Size)); + simple_copy(multimap.begin(), multimap.end(), out_pairs); + simple_mutate_map(multimap.begin(), multimap.end(), MutationFunc()); + VERIFY (validate_multi_associative(multimap, arr, arr + Size, MutationFunc())); + simple_copy(multimap.begin(), multimap.end(), out_pairs_mut); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_multi_associative(reference_multimap, + out_pairs, out_pairs + Size)); + simple_mutate_map(reference_multimap.begin(), reference_multimap.end(), MutationFunc()); + VERIFY_NON_TARGET (validate_multi_associative(reference_multimap, + out_pairs_mut, out_pairs_mut + Size)); + return true; +} + +template<typename T, std::size_t Size> +bool unordered_multiset_test(const T (&arr)[Size]) +{ + std::unordered_multiset<T> reference_multiset(arr, arr + Size); + bool ok; + T out_arr[Size]; + #pragma omp target map(from: ok, out_arr[:Size]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::unordered_multiset<T> set(arr, arr + Size); + VERIFY (validate_multi_associative(set, arr, arr + Size)); + simple_copy(set.begin(), set.end(), out_arr); + /* Sets can't be mutated, we could create another set with mutated + but it gets a little annoying and probably isn't an interesting test. */ + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (validate_multi_associative(reference_multiset, + out_arr, out_arr + Size)); + return true; +} + +#else +template<typename, typename T, std::size_t Size> bool array_test(const T (&arr)[Size]) { return true; } +template<typename, typename T, std::size_t Size> bool forward_list_test(const T (&arr)[Size]) { return true; } +template<typename, typename T, std::size_t Size> bool unordered_map_test(const T (&arr)[Size]) { return true; } +template<typename T, std::size_t Size> bool unordered_set_test(const T (&arr)[Size]) { return true; } +template<typename, typename T, std::size_t Size> bool unordered_multimap_test(const T (&arr)[Size]) { return true; } +template<typename T, std::size_t Size> bool unordered_multiset_test(const T (&arr)[Size]) { return true; } +#endif + +/* This clamps to the maximum value to guard against overflowing, + assuming std::numeric_limits is specialized for T. */ +struct multiply_by_2 +{ + template<typename T> + typename enable_if<std::numeric_limits<T>::is_specialized, T>::type + operator()(T arg) const BL_NOEXCEPT { + if (arg < static_cast<T>(0)) + { + if (std::numeric_limits<T>::min() / static_cast<T>(2) >= arg) + return std::numeric_limits<T>::min(); + } + else + { + if (std::numeric_limits<T>::max() / static_cast<T>(2) <= arg) + return std::numeric_limits<T>::max(); + } + return arg * 2; + } + template<typename T> + typename enable_if<!std::numeric_limits<T>::is_specialized, T>::type + operator()(T arg) const BL_NOEXCEPT { + return arg * 2; + } +}; + +int main() +{ + int data[8] = {0, 1, 2, 3, 4, 5, 6, 7}; + std::pair<int, int> pairs[10] = {std::pair<int, int>( 1, 2), + std::pair<int, int>( 2, 4), + std::pair<int, int>( 3, 6), + std::pair<int, int>( 4, 8), + std::pair<int, int>( 5, 10), + std::pair<int, int>( 6, 12), + std::pair<int, int>( 7, 14), + std::pair<int, int>( 8, 16), + std::pair<int, int>( 9, 18), + std::pair<int, int>(10, 20)}; + const bool vec_res = vector_test<multiply_by_2>(data); + const bool deque_res = deque_test<multiply_by_2>(data); + const bool list_res = list_test<multiply_by_2>(data); + const bool map_res = map_test<multiply_by_2>(pairs); + const bool set_res = set_test(data); + const bool multimap_res = multimap_test<multiply_by_2>(pairs); + const bool multiset_res = multiset_test(data); + const bool array_res = array_test<multiply_by_2>(data); + const bool forward_list_res = forward_list_test<multiply_by_2>(data); + const bool unordered_map_res = unordered_map_test<multiply_by_2>(pairs); + const bool unordered_set_res = unordered_set_test(data); + const bool unordered_multimap_res = unordered_multimap_test<multiply_by_2>(pairs); + const bool unordered_multiset_res = unordered_multiset_test(data); + std::printf("vector : %s\n", vec_res ? "PASS" : "FAIL"); + std::printf("deque : %s\n", deque_res ? "PASS" : "FAIL"); + std::printf("list : %s\n", list_res ? "PASS" : "FAIL"); + std::printf("map : %s\n", map_res ? "PASS" : "FAIL"); + std::printf("set : %s\n", set_res ? "PASS" : "FAIL"); + std::printf("multimap : %s\n", multimap_res ? "PASS" : "FAIL"); + std::printf("multiset : %s\n", multiset_res ? "PASS" : "FAIL"); + std::printf("array : %s\n", array_res ? "PASS" : "FAIL"); + std::printf("forward_list : %s\n", forward_list_res ? "PASS" : "FAIL"); + std::printf("unordered_map : %s\n", unordered_map_res ? "PASS" : "FAIL"); + std::printf("unordered_set : %s\n", unordered_set_res ? "PASS" : "FAIL"); + std::printf("unordered_multimap: %s\n", unordered_multimap_res ? "PASS" : "FAIL"); + std::printf("unordered_multiset: %s\n", unordered_multiset_res ? "PASS" : "FAIL"); + const bool ok = vec_res + && deque_res + && list_res + && map_res + && set_res + && multimap_res + && multiset_res + && array_res + && forward_list_res + && unordered_map_res + && unordered_set_res + && unordered_multimap_res + && unordered_multiset_res; + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-2000.C b/libgomp/testsuite/libgomp.c++/target-flex-2000.C new file mode 100644 index 0000000..688c014 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-2000.C @@ -0,0 +1,32 @@ +/* Tiny tuple test. */ + +#include <tuple> + +#include "target-flex-common.h" + +bool test(int arg) +{ + bool ok; + int out; + std::tuple tup = {'a', arg, 3.14f}; + #pragma omp target map(from: ok, out) map(to: tup) + { + bool inner_ok = true; + { + VERIFY (std::get<0>(tup) == 'a'); + out = std::get<1>(tup); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (out == arg); + return true; +} + +int main() +{ + volatile int arg = 42u; + return test(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-2001.C b/libgomp/testsuite/libgomp.c++/target-flex-2001.C new file mode 100644 index 0000000..f1a6c12 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-2001.C @@ -0,0 +1,61 @@ +/* { dg-additional-options "-std=c++20" } */ + +/* Functional */ + +#include <functional> +#include <utility> + +#include "target-flex-common.h" + +template<typename T,typename Fn> +auto invoke_unary(T&& a, Fn&& fn) noexcept +{ + return std::invoke(std::forward<Fn>(fn), + std::forward<T>(a)); +} + +template<typename T, typename U, typename Fn> +auto invoke_binary(T&& a, U&& b, Fn&& fn) noexcept +{ + return std::invoke(std::forward<Fn>(fn), + std::forward<T>(a), + std::forward<U>(b)); +} + +bool test(unsigned arg) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arg) + { + bool inner_ok = true; + { + VERIFY (std::plus{}(arg, 2) == arg + 2); + auto bound_plus_arg = std::bind_front(std::plus{}, arg); + VERIFY (bound_plus_arg(10) == arg + 10); + VERIFY (bound_plus_arg(20) == arg + 20); + + VERIFY (std::not_fn(std::not_equal_to{})(arg, arg)); + VERIFY (invoke_binary(arg, arg, std::not_fn(std::not_equal_to{}))); + auto bound_equals_arg = std::bind_front(std::not_fn(std::not_equal_to{}), arg); + VERIFY (bound_equals_arg(arg)); + VERIFY (std::not_fn(bound_equals_arg)(arg + 1)); + VERIFY (invoke_unary(arg, bound_equals_arg)); + + VERIFY (std::not_fn(std::ranges::not_equal_to{})(arg, arg)); + VERIFY (invoke_binary(arg, arg, std::not_fn(std::ranges::not_equal_to{}))); + auto bound_ranges_equals_arg = std::bind_front(std::not_fn(std::ranges::not_equal_to{}), arg); + VERIFY (bound_ranges_equals_arg(arg)); + VERIFY (std::not_fn(bound_ranges_equals_arg)(arg + 1)); + VERIFY (invoke_unary(arg, bound_ranges_equals_arg)); + } + end: + ok = inner_ok; + } + return ok; +} + +int main() +{ + volatile unsigned arg = 42u; + return test(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-2002.C b/libgomp/testsuite/libgomp.c++/target-flex-2002.C new file mode 100644 index 0000000..f738806 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-2002.C @@ -0,0 +1,97 @@ +/* { dg-additional-options "-std=c++23" } */ + +/* expected/optional */ + +#include <optional> +#include <expected> + +#include "target-flex-common.h" + +std::optional<unsigned> make_optional(bool b, unsigned arg = 0u) noexcept +{ + if (!b) + return std::nullopt; + return {arg}; +} + +bool test_optional(unsigned arg) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arg) + { + bool inner_ok = true; + { + auto null_opt = make_optional(false); + VERIFY (!null_opt); + VERIFY (!null_opt.has_value()); + VERIFY (null_opt.value_or(arg * 2u) == arg * 2u); + VERIFY (null_opt.or_else([&](){ return std::optional<unsigned>{arg}; }) + .transform([](int a){ return a * 2u; }) + .value_or(0) == arg * 2u); + + auto opt = make_optional(true, arg); + VERIFY (opt); + VERIFY (opt.has_value()); + VERIFY (opt.value() == arg); + VERIFY (*opt == arg); + VERIFY (opt.value_or(arg + 42) == arg); + VERIFY (opt.or_else([&](){ return std::optional<unsigned>{arg + 42}; }) + .transform([](int a){ return a * 2u; }) + .value_or(0) == arg * 2u); + } + end: + ok = inner_ok; + } + return ok; +} + +struct my_error +{ + int _e; +}; + +std::expected<unsigned, my_error> make_expected(bool b, unsigned arg = 0u) noexcept +{ + if (!b) + return std::unexpected{my_error{-1}}; + return {arg}; +} + +bool test_expected(unsigned arg) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arg) + { + bool inner_ok = true; + { + auto unexpected = make_expected(false); + VERIFY (!unexpected); + VERIFY (!unexpected.has_value()); + VERIFY (unexpected.error()._e == -1); + VERIFY (unexpected.value_or(arg * 2u) == arg * 2u); + VERIFY (unexpected.or_else([&](my_error e){ return std::expected<unsigned, my_error>{arg}; }) + .transform([](int a){ return a * 2u; }) + .value_or(0) == arg * 2u); + + auto expected = make_expected(true, arg); + VERIFY (expected); + VERIFY (expected.has_value()); + VERIFY (expected.value() == arg); + VERIFY (*expected == arg); + VERIFY (expected.value_or(arg + 42) == arg); + VERIFY (expected.or_else([&](my_error e){ return std::expected<unsigned, my_error>{std::unexpected{e}}; }) + .transform([](int a){ return a * 2u; }) + .value_or(0) == arg * 2u); + } + end: + ok = inner_ok; + } + return ok; +} + +int main() +{ + volatile unsigned arg = 42; + return test_optional(arg) + && test_expected(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-2003.C b/libgomp/testsuite/libgomp.c++/target-flex-2003.C new file mode 100644 index 0000000..8e8ca8e --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-2003.C @@ -0,0 +1,176 @@ +/* { dg-additional-options "-std=c++20" } */ + +/* bit_cast and memcpy */ + +#include <bit> +#include <cstring> + +#include "target-flex-common.h" + +struct S0 +{ + int _v0; + char _v1; + long long _v2; +}; + +struct S1 +{ + int _v0; + char _v1; + long long _v2; +}; + +bool test_bit_cast(int arg) +{ + bool ok; + S1 s1_out; + #pragma omp target map(from: ok, s1_out) map(to: arg) + { + bool inner_ok = true; + { + long long v = static_cast<long long>(arg + 42ll); + S0 s = {arg, 'a', v}; + VERIFY (std::bit_cast<S1>(s)._v0 == arg); + VERIFY (std::bit_cast<S1>(s)._v1 == 'a'); + VERIFY (std::bit_cast<S1>(s)._v2 == v); + s1_out = std::bit_cast<S1>(s); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + long long v = static_cast<long long>(arg + 42ll); + VERIFY_NON_TARGET (std::bit_cast<S0>(s1_out)._v0 == arg); + VERIFY_NON_TARGET (std::bit_cast<S0>(s1_out)._v1 == 'a'); + VERIFY_NON_TARGET (std::bit_cast<S0>(s1_out)._v2 == v); + return true; +} + + +struct OutStruct +{ + std::size_t _id; + void *_next; +}; + +struct Extendable1 +{ + std::size_t _id; + void *_next; + int _v; +}; + +struct Extendable2 +{ + std::size_t _id; + void *_next; + char _str[256]; +}; + +struct Extendable3 +{ + std::size_t _id; + void *_next; + const int *_nums; + std::size_t _size; +}; + +struct ExtendableUnknown +{ + std::size_t _id; + void *_next; +}; + +template<typename To, std::size_t Id> +To *get_extendable(void *p) +{ + while (p != nullptr) + { + OutStruct out; + std::memcpy(&out, p, sizeof(OutStruct)); + if (out._id == Id) + return static_cast<To *>(p); + p = out._next; + } + return nullptr; +} + +bool test_memcpy(int arg, const int *nums, std::size_t nums_size) +{ + bool ok; + Extendable2 e2_out; + #pragma omp target map(from: ok, e2_out) map(to: arg, nums[:nums_size], nums_size) + { + bool inner_ok = true; + { + Extendable3 e3 = {3u, nullptr, nums, nums_size}; + ExtendableUnknown u1 = {100u, &e3}; + Extendable2 e2 = {2u, &u1, {'H', 'e', 'l', 'l', 'o', '!', '\000'}}; + ExtendableUnknown u2 = {101u, &e2}; + ExtendableUnknown u3 = {102u, &u2}; + ExtendableUnknown u4 = {142u, &u3}; + Extendable1 e1 = {1u, &u4, arg}; + + void *p = &e1; + while (p != nullptr) + { + /* You can always cast a pointer to a struct to a pointer to + the type of it's first member. */ + switch (*static_cast<std::size_t *>(p)) + { + case 1: + { + Extendable1 *e1_p = static_cast<Extendable1 *>(p); + p = e1_p->_next; + VERIFY (e1_p->_v == arg); + break; + } + case 2: + { + Extendable2 *e2_p = static_cast<Extendable2 *>(p); + p = e2_p->_next; + VERIFY (std::strcmp(e2_p->_str, "Hello!") == 0); + break; + } + case 3: + { + Extendable3 *e3_p = static_cast<Extendable3 *>(p); + p = e3_p->_next; + VERIFY (nums == e3_p->_nums); + VERIFY (nums_size == e3_p->_size); + break; + } + default: + { + /* Casting to a pointer to OutStruct invokes undefined + behavior though, memcpy is required to extract the _next + member. */ + OutStruct out; + std::memcpy(&out, p, sizeof(OutStruct)); + p = out._next; + } + } + } + Extendable2 *e2_p = get_extendable<Extendable2, 2u>(&e1); + VERIFY (e2_p != nullptr); + e2_out = *e2_p; + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (e2_out._id == 2u); + VERIFY_NON_TARGET (std::strcmp(e2_out._str, "Hello!") == 0); + return true; +} + +int main() +{ + volatile int arg = 42; + int arr[8] = {0, 1, 2, 3, 4, 5, 6, 7}; + return test_bit_cast(arg) + && test_memcpy(arg, arr, 8) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-30.C b/libgomp/testsuite/libgomp.c++/target-flex-30.C new file mode 100644 index 0000000..c66075b --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-30.C @@ -0,0 +1,51 @@ +/* std::initializer_list in target region. */ + +#include <initializer_list> +#include <array> + +#include "target-flex-common.h" + +bool test_initializer_list(int arg) +{ + static constexpr std::size_t out_arr_size = 7; + int out_arr[out_arr_size]; + bool ok; + #pragma omp target map(from: ok, out_arr[:out_arr_size]) map(to: arg) + { + bool inner_ok = true; + { + auto il = {0, 1, 2, 3, 4, 5, arg}; + + int sum = 0; + for (auto const& e : il) + sum += e; + VERIFY (sum == 0 + 1 + 2 + 3 + 4 + 5 + arg); + + auto* out_it = out_arr; + const auto* const out_end = out_arr + out_arr_size; + for (auto const& e : il) + { + VERIFY (out_it != out_end); + *out_it = e; + ++out_it; + } + } + end: + ok = inner_ok; + } + if (!ok) + return false; + + std::array<int, out_arr_size> reference_array = {0, 1, 2, 3, 4, 5, arg}; + const auto *out_arr_it = out_arr; + for (auto const& e : reference_array) + VERIFY_NON_TARGET (e == *(out_arr_it++)); + + return true; +} + +int main() +{ + volatile int arg = 42; + return test_initializer_list(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-300.C b/libgomp/testsuite/libgomp.c++/target-flex-300.C new file mode 100644 index 0000000..329a189 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-300.C @@ -0,0 +1,51 @@ +/* { dg-additional-options -std=c++23 } */ + +/* numerics */ + +#include <algorithm> +#include <numeric> +#include <ranges> +#include <span> +#include <vector> + +//TODO PR120454 "C++ constexpr vs. OpenMP implicit mapping" +#pragma omp declare target(std::ranges::all_of, std::ranges::iota) + +#include "target-flex-common.h" + +namespace stdr = std::ranges; + +bool test(std::size_t arg) +{ + bool ok; + int midpoint_out; + std::vector<int> vec(arg); + int *data = vec.data(); + std::size_t size = vec.size(); + #pragma omp target defaultmap(none) map(from: ok, midpoint_out) map(tofrom: data[:size]) map(to: arg, size) + /* <https://baylibre.slack.com/archives/C06TTV7HMMG/p1748508583437829> + { dg-bogus {sorry, unimplemented: unsupported map expression '<lambda closure object>.*} TODO { xfail *-*-* } .-2 } */ + { + std::span span = {data, size}; + bool inner_ok = true; + { + VERIFY (stdr::all_of(span, [](int v){ return v == int{}; })); + stdr::iota(span, 0); + midpoint_out = *std::midpoint(span.data(), span.data() + span.size()); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (stdr::equal(vec, std::views::iota(0, static_cast<int>(vec.size())))); + VERIFY_NON_TARGET (*std::midpoint(vec.data(), vec.data() + vec.size()) + == midpoint_out); + return true; +} + +int main() +{ + volatile std::size_t arg = 42; + return test(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-31.C b/libgomp/testsuite/libgomp.c++/target-flex-31.C new file mode 100644 index 0000000..adaf18f --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-31.C @@ -0,0 +1,80 @@ +/* std::initializer_list in target region. */ + +#include <initializer_list> + +#include "target-flex-common.h" + +struct S0 +{ + int _v; + S0(std::initializer_list<int> il) + : _v(0) + { + for (auto const& e : il) + _v += e; + } +}; + +struct S1 +{ + int _v; + template<typename T> + S1(std::initializer_list<T> il) + : _v(0) + { + for (auto const& e : il) + _v += e; + } +}; + +template<typename T> +struct S2 +{ + T _v; + S2(std::initializer_list<T> il) + : _v(0) + { + for (auto const& e : il) + _v += e; + } +}; + +#if __cplusplus >= 201703L +template<typename T> +S2(std::initializer_list<T>) -> S2<T>; +#endif + +bool test_initializer_list(int arg) +{ + bool ok; + #pragma omp target map(from: ok) map(to: arg) + { + bool inner_ok = true; + { + static constexpr int partial_sum = 0 + 1 + 2 + 3 + 4 + 5; + + S0 s0{0, 1, 2, 3, 4, 5, arg}; + VERIFY (s0._v == partial_sum + arg); + + S1 s1{0, 1, 2, 3, 4, 5, arg}; + VERIFY (s1._v == partial_sum + arg); + + S2<int> s2{0, 1, 2, 3, 4, 5, arg}; + VERIFY (s2._v == partial_sum + arg); + + #if __cplusplus >= 201703L + S2 s2_ctad{0, 1, 2, 3, 4, 5, arg}; + VERIFY (s2_ctad._v == partial_sum + arg); + #endif + } + end: + ok = inner_ok; + } + return ok; +} + +int main() +{ + volatile int arg = 42; + return test_initializer_list(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-32.C b/libgomp/testsuite/libgomp.c++/target-flex-32.C new file mode 100644 index 0000000..7f74401a --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-32.C @@ -0,0 +1,50 @@ +/* std::initializer_list constructor of std::vector (explicit template arg) */ + +#include <vector> +#include <array> + +#include "target-flex-common.h" + +bool test_initializer_list(int arg) +{ + static constexpr std::size_t out_arr_size = 7; + int out_arr[out_arr_size]; + bool ok; + #pragma omp target map(from: ok, out_arr[:out_arr_size]) map(to: arg) + { + bool inner_ok = true; + { + std::vector<int> vec{0, 1, 2, 3, 4, 5, arg}; + int sum = 0; + for (auto const& e : vec) + sum += e; + VERIFY (sum == 0 + 1 + 2 + 3 + 4 + 5 + arg); + + auto* out_it = out_arr; + const auto* const out_end = out_arr + out_arr_size; + for (auto const& e : vec) + { + VERIFY (out_it != out_end); + *out_it = e; + ++out_it; + } + } + end: + ok = inner_ok; + } + if (!ok) + return false; + + std::array<int, out_arr_size> reference_array = {0, 1, 2, 3, 4, 5, arg}; + const auto *out_arr_it = out_arr; + for (auto const& e : reference_array) + VERIFY_NON_TARGET (e == *(out_arr_it++)); + + return true; +} + +int main() +{ + volatile int arg = 42; + return test_initializer_list(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-33.C b/libgomp/testsuite/libgomp.c++/target-flex-33.C new file mode 100644 index 0000000..bb8a39b --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-33.C @@ -0,0 +1,52 @@ +/* { dg-additional-options "-std=c++17" } */ + +/* deduced std::initializer_list constructor of std::vector (CTAD) */ + +#include <vector> +#include <array> + +#include "target-flex-common.h" + +bool test_initializer_list(int arg) +{ + static constexpr std::size_t out_arr_size = 7; + int out_arr[out_arr_size]; + bool ok; + #pragma omp target map(from: ok, out_arr[:out_arr_size]) map(to: arg) + { + bool inner_ok = true; + { + std::vector vec{0, 1, 2, 3, 4, 5, arg}; + int sum = 0; + for (auto const& e : vec) + sum += e; + VERIFY (sum == 0 + 1 + 2 + 3 + 4 + 5 + arg); + + auto* out_it = out_arr; + const auto* const out_end = out_arr + out_arr_size; + for (auto const& e : vec) + { + VERIFY (out_it != out_end); + *out_it = e; + ++out_it; + } + } + end: + ok = inner_ok; + } + if (!ok) + return false; + + std::array<int, out_arr_size> reference_array = {0, 1, 2, 3, 4, 5, arg}; + const auto *out_arr_it = out_arr; + for (auto const& e : reference_array) + VERIFY_NON_TARGET (e == *(out_arr_it++)); + + return true; +} + +int main() +{ + volatile int arg = 42; + return test_initializer_list(arg) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-41.C b/libgomp/testsuite/libgomp.c++/target-flex-41.C new file mode 100644 index 0000000..4d36341 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-41.C @@ -0,0 +1,94 @@ +/* { dg-additional-options "-std=c++20" } */ + +/* <iterator> c++20 */ + +/* std::common_iterator uses std::variant. */ + +#include <vector> +#include <iterator> +#include <span> + +//TODO PR120454 "C++ constexpr vs. OpenMP implicit mapping" +#pragma omp declare target(std::ranges::distance, std::ranges::next) + +#include "target-flex-common.h" + +namespace stdr = std::ranges; + +template<typename It0, typename It1> +bool simple_equal(const It0 begin0, const It0 end0, + const It1 begin1, const It1 end1) BL_NOEXCEPT +{ + It0 it0 = begin0; + It1 it1 = begin1; + for (; it0 != end0; ++it0, ++it1) + if (it1 == end1 || *it0 != *it1) + return false; + return true; +} + +template<typename It, typename OutIt> +void simple_copy(const It begin, const It end, OutIt out) BL_NOEXCEPT +{ + for (It it = begin; it != end; ++it, ++out) + *out = *it; +} + +template<typename T, std::size_t Size> +bool test(const T (&arr)[Size]) +{ + bool ok; + T out_rev_arr[Size]; + T out_fwd_arr[Size]; + T out_first_half_arr[Size / 2]; + #pragma omp target defaultmap(none) \ + map(from: ok, out_rev_arr[:Size], out_fwd_arr[:Size], \ + out_first_half_arr[:Size / 2]) \ + map(to: arr[:Size]) + { + bool inner_ok = true; + { + std::span<const T> span = {arr, Size}; + std::vector<T> rev_vec(std::reverse_iterator{span.end()}, + std::reverse_iterator{span.begin()}); + VERIFY (std::distance(span.begin(), span.end()) + == std::distance(rev_vec.begin(), rev_vec.end())); + VERIFY (stdr::distance(span.begin(), span.end()) + == stdr::distance(rev_vec.begin(), rev_vec.end())); + VERIFY (stdr::distance(span) == stdr::distance(rev_vec)); + VERIFY (simple_equal(span.begin(), span.end(), + std::reverse_iterator{rev_vec.end()}, + std::reverse_iterator{rev_vec.begin()})); + simple_copy(rev_vec.begin(), rev_vec.end(), out_rev_arr); + simple_copy(std::reverse_iterator{rev_vec.end()}, + std::reverse_iterator{rev_vec.begin()}, + out_fwd_arr); + using counted_iter = std::counted_iterator<decltype(span.begin())>; + using common_iter = std::common_iterator<counted_iter, + std::default_sentinel_t>; + std::vector<T> front_half; + simple_copy(common_iter{counted_iter{span.begin(), Size / 2}}, + common_iter{std::default_sentinel}, + std::back_insert_iterator{front_half}); + VERIFY (simple_equal(span.begin(), stdr::next(span.begin(), Size / 2), + front_half.begin(), front_half.end())); + simple_copy(front_half.begin(), front_half.end(), out_first_half_arr); + } + end: + ok = inner_ok; + } + VERIFY_NON_TARGET (simple_equal(std::reverse_iterator{arr + Size}, + std::reverse_iterator{arr}, + out_rev_arr, out_rev_arr + Size)); + VERIFY_NON_TARGET (simple_equal(arr, arr + Size, + out_fwd_arr, out_fwd_arr + Size)); + VERIFY_NON_TARGET (simple_equal(arr, arr + Size / 2, + out_first_half_arr, out_first_half_arr + Size / 2)); + return ok; +} + +int main() +{ + int arr[] = {0, 1, 2, 3, 4, 5, 6, 7}; + return test(arr) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-60.C b/libgomp/testsuite/libgomp.c++/target-flex-60.C new file mode 100644 index 0000000..393bb3c --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-60.C @@ -0,0 +1,48 @@ +/* algorithms pre c++20 */ + +#include <algorithm> +#include <vector> + +#include "target-flex-common.h" + +template<typename T, std::size_t Size> +bool test(const T (&arr)[Size]) +{ + bool ok; + T out_2x_arr[Size]; + T out_shifted_arr[Size]; + #pragma omp target map(from: ok, out_2x_arr[:Size], out_shifted_arr[:Size]) \ + map(to: arr[:Size]) + /* <https://baylibre.slack.com/archives/C06TTV7HMMG/p1748508583437829> + { dg-bogus {sorry, unimplemented: unsupported map expression '<lambda closure object>.*} TODO { xfail *-*-* } .-3 } */ + { + std::vector<T> vec(Size); + std::vector<T> mutated(Size); + bool inner_ok = true; + { + std::copy(arr, arr + Size, vec.begin()); + VERIFY (std::equal(arr, arr + Size, vec.begin())); + std::transform(vec.begin(), vec.end(), mutated.begin(), + [](const T& v){ return v * 2; }); + std::copy(mutated.begin(), mutated.end(), out_2x_arr); + std::rotate(vec.begin(), std::next(vec.begin(), Size / 2), vec.end()); + std::copy(vec.begin(), vec.end(), out_shifted_arr); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (std::equal(arr, arr + Size, out_2x_arr, + [](const T& a, const T& b){ return a * 2 == b; })); + std::vector<T> shifted(arr, arr + Size); + std::rotate(shifted.begin(), std::next(shifted.begin(), Size / 2), shifted.end()); + VERIFY_NON_TARGET (std::equal(out_shifted_arr, out_shifted_arr + Size, shifted.begin())); + return true; +} + +int main() +{ + int arr[] = {0, 1, 2, 3, 4, 5, 6, 7}; + return test(arr) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-61.C b/libgomp/testsuite/libgomp.c++/target-flex-61.C new file mode 100644 index 0000000..e06133a --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-61.C @@ -0,0 +1,56 @@ +/* { dg-additional-options "-std=c++20" } */ + +/* ranged algorithms c++20 */ + +#include <algorithm> +#include <ranges> +#include <vector> + +//TODO PR120454 "C++ constexpr vs. OpenMP implicit mapping" +#pragma omp declare target(std::ranges::copy, std::ranges::equal, std::ranges::rotate, std::ranges::transform) + +#include "target-flex-common.h" + +namespace stdr = std::ranges; + +template<typename T, std::size_t Size> +bool test(const T (&arr)[Size]) +{ + bool ok; + T out_2x_arr[Size]; + T out_shifted_arr[Size]; + #pragma omp target defaultmap(none) \ + map(from: ok, out_2x_arr[:Size], out_shifted_arr[:Size]) \ + map(to: arr[:Size]) + /* <https://baylibre.slack.com/archives/C06TTV7HMMG/p1748508583437829> + { dg-bogus {sorry, unimplemented: unsupported map expression '<lambda closure object>.*} TODO { xfail *-*-* } .-4 } */ + { + std::vector<T> vec(Size); + std::vector<T> mutated(Size); + bool inner_ok = true; + { + stdr::copy(arr, vec.begin()); + VERIFY (stdr::equal(arr, vec)); + stdr::transform(vec, mutated.begin(), + [](const T& v){ return v * 2; }); + stdr::copy(mutated, out_2x_arr); + stdr::rotate(vec, std::next(vec.begin(), Size / 2)); + stdr::copy(vec, out_shifted_arr); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (stdr::equal(arr, out_2x_arr, stdr::equal_to{}, [](const T& v){ return v * 2; })); + std::vector<T> shifted(arr, arr + Size); + stdr::rotate(shifted, std::next(shifted.begin(), Size / 2)); + VERIFY_NON_TARGET (stdr::equal(out_shifted_arr, shifted)); + return true; +} + +int main() +{ + int arr[] = {0, 1, 2, 3, 4, 5, 6, 7}; + return test(arr) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-62.C b/libgomp/testsuite/libgomp.c++/target-flex-62.C new file mode 100644 index 0000000..2e74b20 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-62.C @@ -0,0 +1,52 @@ +/* { dg-additional-options -std=c++23 } */ + +/* std::views stuff. Also tests std::tuple with std::views::zip. */ + +#include <algorithm> +#include <ranges> +#include <span> + +//TODO PR120454 "C++ constexpr vs. OpenMP implicit mapping" +#pragma omp declare target(std::ranges::all_of, std::ranges::equal, std::ranges::fold_left, std::views::reverse, std::views::zip) + +#include "target-flex-common.h" + +namespace stdr = std::ranges; +namespace stdv = std::views; + +bool f() +{ + const int arr_fwd[8] = {0, 1, 2, 3, 4, 5, 6, 7}; + const int arr_rev[8] = {7, 6, 5, 4, 3, 2, 1, 0}; + + bool ok; + #pragma omp target defaultmap(none) map(from: ok) map(to: arr_fwd[:8], arr_rev[:8]) + /* <https://baylibre.slack.com/archives/C06TTV7HMMG/p1748508583437829> + { dg-bogus {sorry, unimplemented: unsupported map expression '<lambda closure object>.*} TODO { xfail *-*-* } .-2 } */ + { + std::span<const int> fwd = {arr_fwd, 8}; + std::span<const int> rev = {arr_rev, 8}; + bool inner_ok = true; + { + VERIFY(stdr::equal(fwd, rev | stdv::reverse)); + VERIFY(stdr::equal(fwd | stdv::drop(4) | stdv::reverse, + rev | stdv::take(4))); + for (auto [first, second] : stdv::zip(fwd, rev)) + VERIFY(first + second == 7); + auto plus = [](int a, int b){ return a + b; }; + auto is_even = [](int v){ return v % 2 == 0; }; + VERIFY(stdr::fold_left(fwd | stdv::filter(is_even), 0, plus) + == 12); + VERIFY(stdr::all_of(fwd | stdv::transform([](int v){ return v * 2; }), + is_even)); + } + end: + ok = inner_ok; + } + return ok; +} + +int main() +{ + return f() ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-70.C b/libgomp/testsuite/libgomp.c++/target-flex-70.C new file mode 100644 index 0000000..9e9383d --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-70.C @@ -0,0 +1,26 @@ +/* CTAD in target regions. */ + +template<typename T> +struct S +{ + T _v; +}; + +template<typename T> +S(T) -> S<T>; + +bool f() +{ + bool ok; + #pragma omp target map(from: ok) + { + S s{42}; + ok = s._v == 42; + } + return ok; +} + +int main() +{ + return f() ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-80.C b/libgomp/testsuite/libgomp.c++/target-flex-80.C new file mode 100644 index 0000000..f41a1bb --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-80.C @@ -0,0 +1,49 @@ +// { dg-additional-options "-std=c++20" } + +/* std::span */ + +#include <span> + +#include "target-flex-common.h" + +template<typename It0, typename It1> +bool simple_equal(It0 it0, const It0 end0, + It1 it1, const It1 end1) noexcept +{ + for (; it0 != end0; ++it0, ++it1) + if (it1 == end1 || *it0 != *it1) + return false; + return true; +} + +template<typename T, std::size_t Size> +bool test(const T (&arr)[Size]) +{ + bool ok; + T out_arr[Size]; + #pragma omp target map(from: ok) map(to: arr[:Size]) + { + std::span span = {arr, Size}; + bool inner_ok = true; + { + VERIFY (!span.empty()); + VERIFY (span.size() == Size); + auto out_it = out_arr; + for (auto elem : span) + *(out_it++) = elem; + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (simple_equal(arr, arr + Size, + out_arr, out_arr + Size)); + return true; +} + +int main() +{ + int arr[8] = {0, 1, 2, 3, 4, 5, 6, 7}; + return test(arr) ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-81.C b/libgomp/testsuite/libgomp.c++/target-flex-81.C new file mode 100644 index 0000000..950c122 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-81.C @@ -0,0 +1,77 @@ +/* { dg-additional-options "-std=c++20" } */ + +#include <ranges> +#include <span> +#include <type_traits> +#include <vector> + +#include "target-flex-common.h" + +namespace stdr = std::ranges; + +template<typename It0, typename It1> +bool simple_equal(It0 it0, const It0 end0, + It1 it1, const It1 end1) noexcept +{ + for (; it0 != end0; ++it0, ++it1) + if (it1 == end1 || *it0 != *it1) + return false; + return true; +} + +template<typename Rn0, typename Rn1> +bool simple_equal(Rn0&& rn0, Rn1&& rn1) noexcept +{ + return simple_equal(stdr::begin(rn0), stdr::end(rn0), + stdr::begin(rn1), stdr::end(rn1)); +} + +template<typename Rn> +bool test(Rn&& range) +{ + using value_type = stdr::range_value_t<std::remove_cvref_t<Rn>>; + std::vector<value_type> vec = {stdr::begin(range), stdr::end(range)}; + value_type *data = vec.data(); + std::size_t size = vec.size(); + bool ok; + #pragma omp target map(from: ok) map(tofrom: data[:size]) map(to: size) + /* <https://baylibre.slack.com/archives/C06TTV7HMMG/p1748508583437829> + { dg-bogus {sorry, unimplemented: unsupported map expression '<lambda closure object>.*} TODO { xfail *-*-* } .-2 } */ + { + std::vector<value_type> orig = {data, data + size}; + std::span<value_type> span = {data, size}; + bool inner_ok = true; + { + auto mul_by_2 = [](const value_type& v){ return v * 2; }; + VERIFY (simple_equal(orig, span)); + for (auto& elem : span) + elem = mul_by_2(elem); + VERIFY (simple_equal(orig | std::views::transform(mul_by_2), span)); + } + end: + ok = inner_ok; + } + if (!ok) + return false; + auto mul_by_2 = [](const value_type& v){ return v * 2; }; + VERIFY_NON_TARGET (simple_equal(range | std::views::transform(mul_by_2), vec)); + return true; +} + +struct my_int +{ + int _v; + bool operator==(my_int const&) const = default; + my_int operator*(int rhs) const noexcept { + return {_v * rhs}; + } +}; + +int main() +{ + std::vector<int> ints = {1, 2, 3, 4, 5}; + const bool ints_res = test(ints); + std::vector<my_int> my_ints = {my_int{1}, my_int{2}, my_int{3}, my_int{4}, my_int{5}}; + const bool my_ints_res = test(my_ints); + return ints_res && my_ints_res ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-90.C b/libgomp/testsuite/libgomp.c++/target-flex-90.C new file mode 100644 index 0000000..b3f1197 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-90.C @@ -0,0 +1,107 @@ +/* structured bindings */ + +#include <array> +#include <tuple> + +#include "target-flex-common.h" + +template<typename Array, typename Tuple, typename Struct> +bool test(Array array, Tuple tuple, Struct s) +{ + bool ok; + auto array_2nd_in = std::get<2>(array); + auto tuple_2nd_in = std::get<2>(tuple); + auto s_2nd_in = s._2; + decltype(array_2nd_in) array_2nd_out_0; + decltype(tuple_2nd_in) tuple_2nd_out_0; + decltype(s_2nd_in) s_2nd_out_0; + decltype(array_2nd_in) array_2nd_out_1; + decltype(tuple_2nd_in) tuple_2nd_out_1; + decltype(s_2nd_in) s_2nd_out_1; + decltype(array_2nd_in) array_2nd_out_2; + decltype(tuple_2nd_in) tuple_2nd_out_2; + decltype(s_2nd_in) s_2nd_out_2; + #pragma omp target map(from: ok, \ + array_2nd_out_0, tuple_2nd_out_0, s_2nd_out_0, \ + array_2nd_out_1, tuple_2nd_out_1, s_2nd_out_1, \ + array_2nd_out_2, tuple_2nd_out_2, s_2nd_out_2) \ + map(to: array_2nd_in, tuple_2nd_in, s_2nd_in, array, tuple, s) + { + bool inner_ok = true; + { + { + auto [array_0th, array_1st, array_2nd] = array; + VERIFY (array_2nd_in == array_2nd); + VERIFY (std::get<2>(array) == array_2nd); + array_2nd_out_0 = array_2nd; + auto [tuple_0th, tuple_1st, tuple_2nd] = tuple; + VERIFY (tuple_2nd_in == tuple_2nd); + VERIFY (std::get<2>(tuple) == tuple_2nd); + tuple_2nd_out_0 = tuple_2nd; + auto [s_0th, s_1st, s_2nd] = s; + VERIFY (s_2nd_in == s_2nd); + VERIFY (s._2 == s_2nd); + s_2nd_out_0 = s_2nd; + } + { + auto& [array_0th, array_1st, array_2nd] = array; + VERIFY (array_2nd_in == array_2nd); + VERIFY (std::get<2>(array) == array_2nd); + array_2nd_out_1 = array_2nd; + auto& [tuple_0th, tuple_1st, tuple_2nd] = tuple; + VERIFY (tuple_2nd_in == tuple_2nd); + VERIFY (std::get<2>(tuple) == tuple_2nd); + tuple_2nd_out_1 = tuple_2nd; + auto& [s_0th, s_1st, s_2nd] = s; + VERIFY (s_2nd_in == s_2nd); + VERIFY (s._2 == s_2nd); + s_2nd_out_1 = s_2nd; + } + { + const auto& [array_0th, array_1st, array_2nd] = array; + VERIFY (array_2nd_in == array_2nd); + VERIFY (std::get<2>(array) == array_2nd); + array_2nd_out_2 = array_2nd; + const auto& [tuple_0th, tuple_1st, tuple_2nd] = tuple; + VERIFY (tuple_2nd_in == tuple_2nd); + VERIFY (std::get<2>(tuple) == tuple_2nd); + tuple_2nd_out_2 = tuple_2nd; + const auto& [s_0th, s_1st, s_2nd] = s; + VERIFY (s_2nd_in == s_2nd); + VERIFY (s._2 == s_2nd); + s_2nd_out_2 = s_2nd; + } + } + end: + ok = inner_ok; + } + if (!ok) + return false; + VERIFY_NON_TARGET (array_2nd_out_0 == array_2nd_in); + VERIFY_NON_TARGET (tuple_2nd_out_0 == tuple_2nd_in); + VERIFY_NON_TARGET (s_2nd_out_0 == s_2nd_in); + VERIFY_NON_TARGET (array_2nd_out_1 == array_2nd_in); + VERIFY_NON_TARGET (tuple_2nd_out_1 == tuple_2nd_in); + VERIFY_NON_TARGET (s_2nd_out_1 == s_2nd_in); + VERIFY_NON_TARGET (array_2nd_out_2 == array_2nd_in); + VERIFY_NON_TARGET (tuple_2nd_out_2 == tuple_2nd_in); + VERIFY_NON_TARGET (s_2nd_out_2 == s_2nd_in); + + return true; +} + +struct S +{ + char _0; + float _1; + int _2; +}; + +int main() +{ + const bool test_res + = test(std::array{0, 1, 2}, + std::tuple{'a', 3.14f, 42}, + S{'a', 3.14f, 42}); + return test_res ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-flex-common.h b/libgomp/testsuite/libgomp.c++/target-flex-common.h new file mode 100644 index 0000000..14523c4 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-flex-common.h @@ -0,0 +1,40 @@ +#include <cstdio> + +#if __cplusplus >= 201103L + #define BL_NOEXCEPT noexcept +#else + #define BL_NOEXCEPT throw() +#endif + +#if defined __has_builtin +# if __has_builtin (__builtin_LINE) +# define VERIFY_LINE __builtin_LINE () +# endif +#endif +#if !defined VERIFY_LINE +# define VERIFY_LINE __LINE__ +#endif + +/* I'm not a huge fan of macros but in the interest of keeping the code that + isn't being tested as simple as possible, we use them. */ + +#define VERIFY(EXPR) \ + do { \ + if (!(EXPR)) \ + { \ + std::printf("VERIFY ln: %d `" #EXPR "` evaluated to false\n", \ + VERIFY_LINE); \ + inner_ok = false; \ + goto end; \ + } \ + } while (false) + +#define VERIFY_NON_TARGET(EXPR) \ + do { \ + if (!(EXPR)) \ + { \ + std::printf("VERIFY ln: %d `" #EXPR "` evaluated to false\n", \ + VERIFY_LINE); \ + return false; \ + } \ + } while (false) diff --git a/libgomp/testsuite/libgomp.c++/target-std__array-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__array-concurrent-usm.C new file mode 100644 index 0000000..9923783 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__array-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__array-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__array-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__array-concurrent.C new file mode 100644 index 0000000..c42105a --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__array-concurrent.C @@ -0,0 +1,62 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <array> +#include <algorithm> + +#define N 50000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand (); +} + +#pragma omp declare target +bool validate (const std::array<int,N> &arr, int data[]) +{ + for (int i = 0; i < N; ++i) + if (arr[i] != data[i] * data[i]) + return false; + return true; +} +#pragma omp end declare target + +int main (void) +{ + int data[N]; + bool ok; + std::array<int,N> arr; + + srand (time (NULL)); + init (data); + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: arr) +#endif + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&arr) std::array<int,N> (); +#endif + std::copy (data, data + N, arr.begin ()); + } + + #pragma omp target teams distribute parallel for + for (int i = 0; i < N; ++i) + arr[i] *= arr[i]; + + #pragma omp target map (from: ok) + { + ok = validate (arr, data); +#ifndef MEM_SHARED + arr.~array (); +#endif + } + } + + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__bitset-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__bitset-concurrent-usm.C new file mode 100644 index 0000000..9023ef8 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__bitset-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__bitset-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__bitset-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__bitset-concurrent.C new file mode 100644 index 0000000..4fcce93 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__bitset-concurrent.C @@ -0,0 +1,69 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <bitset> +#include <set> +#include <algorithm> + +#define N 4000 +#define MAX 16384 + +void init (int data[]) +{ + std::set<int> _set; + for (int i = 0; i < N; ++i) + { + // Avoid duplicates in data array. + do + data[i] = rand () % MAX; + while (_set.find (data[i]) != _set.end ()); + _set.insert (data[i]); + } +} + +bool validate (int sum, int data[]) +{ + int total = 0; + for (int i = 0; i < N; ++i) + total += data[i]; + return sum == total; +} + +int main (void) +{ + int data[N]; + std::bitset<MAX> _set; + int sum = 0; + + srand (time (NULL)); + init (data); + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: _set) +#endif + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&_set) std::bitset<MAX> (); +#endif + for (int i = 0; i < N; ++i) + _set[data[i]] = true; + } + + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < MAX; ++i) + if (_set[i]) + sum += i; + +#ifndef MEM_SHARED + #pragma omp target + _set.~bitset (); +#endif + } + + bool ok = validate (sum, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__cmath.C b/libgomp/testsuite/libgomp.c++/target-std__cmath.C new file mode 100644 index 0000000..aaf7152 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__cmath.C @@ -0,0 +1,340 @@ +// { dg-do run } +// { dg-additional-options "-std=c++20" } + +#include <cmath> +#include <numbers> + +#define FP_EQUAL(x,y) (std::abs ((x) - (y)) < 1E-6) + +#pragma omp declare target +template<typename T> bool test_basic () +{ + T x = -3.456789; + T y = 1.234567; + T z = 5.678901; + + if (std::abs (x) != -x) + return false; + if (!FP_EQUAL (std::trunc (x / y) * y + std::fmod (x, y), x)) + return false; + if (!FP_EQUAL (x - std::round (x / y) * y, std::remainder (x, y))) + return false; + if (!FP_EQUAL (std::fma (x, y, z), x * y + z)) + return false; + if (std::fmax (x, y) != (x > y ? x : y)) + return false; + if (std::fmin (x, y) != (x < y ? x : y)) + return false; + if (std::fdim (x, y) != std::max(x - y, (T) 0.0)) + return false; + if (std::fdim (y, x) != std::max(y - x, (T) 0.0)) + return false; + return true; +} + +template<typename T> bool test_exp () +{ + T x = -4.567890; + T y = 2.345678; + + if (!FP_EQUAL (std::exp (x), std::pow (std::numbers::e_v<T>, x))) + return false; + if (!FP_EQUAL (std::exp2 (y), std::pow ((T) 2.0, y))) + return false; + if (!FP_EQUAL (std::expm1 (y), std::exp (y) - (T) 1.0)) + return false; + if (!FP_EQUAL (std::log (std::exp (x)), x)) + return false; + if (!FP_EQUAL (std::log10 (std::pow ((T) 10.0, y)), y)) + return false; + if (!FP_EQUAL (std::log2 (std::exp2 (y)), y)) + return false; + if (!FP_EQUAL (std::log1p (std::expm1 (y)), y)) + return false; + return true; +} + +template<typename T> bool test_power () +{ + T x = 7.234251; + T y = 0.340128; + + if (!FP_EQUAL (std::log (std::pow (x, y)) / std::log (x), y)) + return false; + if (!FP_EQUAL (std::sqrt (x) * std::sqrt (x), x)) + return false; + if (!FP_EQUAL (std::cbrt (x) * std::cbrt (x) * std::cbrt (x), x)) + return false; + if (!FP_EQUAL (std::hypot (x, y), std::sqrt (x * x + y * y))) + return false; + return true; +} + +template<typename T> bool test_trig () +{ + T theta = std::numbers::pi / 4; + T phi = std::numbers::pi / 6; + + if (!FP_EQUAL (std::sin (theta), std::sqrt ((T) 2) / 2)) + return false; + if (!FP_EQUAL (std::sin (phi), 0.5)) + return false; + if (!FP_EQUAL (std::cos (theta), std::sqrt ((T) 2) / 2)) + return false; + if (!FP_EQUAL (std::cos (phi), std::sqrt ((T) 3) / 2)) + return false; + if (!FP_EQUAL (std::tan (theta), 1.0)) + return false; + if (!FP_EQUAL (std::tan (phi), std::sqrt ((T) 3) / 3)) + return false; + + T x = 0.33245623; + + if (!FP_EQUAL (std::asin (std::sin (x)), x)) + return false; + if (!FP_EQUAL (std::acos (std::cos (x)), x)) + return false; + if (!FP_EQUAL (std::atan (std::tan (x)), x)) + return false; + if (!FP_EQUAL (std::atan2 (std::sin (x), std::cos (x)), x)) + return false; + return true; +} + +template<typename T> bool test_hyperbolic () +{ + T x = 0.7423532; + + if (!FP_EQUAL (std::sinh (x), (std::exp (x) - std::exp (-x)) / (T) 2.0)) + return false; + if (!FP_EQUAL (std::cosh (x), (std::exp (x) + std::exp (-x)) / (T) 2.0)) + return false; + if (!FP_EQUAL (std::tanh (x), std::sinh (x) / std::cosh (x))) + return false; + if (!FP_EQUAL (std::asinh (std::sinh (x)), x)) + return false; + if (!FP_EQUAL (std::acosh (std::cosh (x)), x)) + return false; + if (!FP_EQUAL (std::atanh (std::tanh (x)), x)) + return false; + return true; +} + +template<typename T> bool test_erf () +{ + if (!FP_EQUAL (std::erf ((T) 0), 0)) + return false; + if (!FP_EQUAL (std::erf ((T) INFINITY), 1)) + return false; + if (!FP_EQUAL (std::erf ((T) -INFINITY), -1)) + return false; + + if (!FP_EQUAL (std::erfc (0), 1)) + return false; + if (!FP_EQUAL (std::erfc ((T) INFINITY), 0)) + return false; + if (!FP_EQUAL (std::erfc ((T) -INFINITY), 2)) + return false; + + return true; +} + +template<typename T> bool test_gamma () +{ + if (!FP_EQUAL (std::tgamma ((T) 5), 4*3*2*1)) + return false; + if (!FP_EQUAL (std::tgamma ((T) 0.5), std::sqrt (std::numbers::pi_v<T>))) + return false; + if (!FP_EQUAL (std::tgamma ((T) -0.5), (T) -2 * std::sqrt (std::numbers::pi_v<T>))) + return false; + if (!FP_EQUAL (std::tgamma ((T) 2.5), (T) 0.75 * std::sqrt (std::numbers::pi_v<T>))) + return false; + if (!FP_EQUAL (std::tgamma ((T) -2.5), (T) -8.0/15 * std::sqrt (std::numbers::pi_v<T>))) + return false; + + if (!FP_EQUAL (std::lgamma ((T) 5), std::log ((T) 4*3*2*1))) + return false; + if (!FP_EQUAL (std::lgamma ((T) 0.5), std::log (std::sqrt (std::numbers::pi_v<T>)))) + return false; + if (!FP_EQUAL (std::lgamma ((T) 2.5), + std::log ((T) 0.75 * std::sqrt (std::numbers::pi_v<T>)))) + return false; + + return true; +} + +template<typename T> bool test_rounding () +{ + T x = -2.5678; + T y = 3.6789; + + if (std::ceil (x) != -2) + return false; + if (std::floor (x) != -3) + return false; + if (std::trunc (x) != -2) + return false; + if (std::round (x) != -3) + return false; + + if (std::ceil (y) != 4) + return false; + if (std::floor (y) != 3) + return false; + if (std::trunc (y) != 3) + return false; + if (std::round (y) != 4) + return false; + + /* Not testing std::rint and std::nearbyint due to dependence on + floating-point environment. */ + + return true; +} + +template<typename T> bool test_fpmanip () +{ + T x = -2.3456789; + T y = 3.6789012; + int exp; + + T mantissa = std::frexp (x, &exp); + if (std::ldexp (mantissa, exp) != x) + return false; + if (std::logb (x) + 1 != exp) + return false; + if (std::ilogb (x) + 1 != exp) + return false; + if (std::scalbn (x, -exp) != mantissa) + return false; + + T next = std::nextafter (x, y); + if (!(next > x && next < y)) + return false; + +#if 0 + /* TODO Due to 'std::nexttoward' using 'long double to', this triggers a + '80-bit-precision floating-point numbers unsupported (mode ‘XF’)' error + with x86_64 host and nvptx, GCN offload compilers, or + '128-bit-precision floating-point numbers unsupported (mode ‘TF’)' error + with powerpc64le host and nvptx offload compiler, for example; + PR71064 'nvptx offloading: "long double" data type'. + It ought to work on systems where the host's 'long double' is the same as + 'double' ('DF'): aarch64, for example? */ + next = std::nexttoward (x, y); + if (!(next > x && next < y)) + return false; +#endif + + if (std::copysign (x, y) != std::abs (x)) + return false; + if (std::copysign (y, x) != -y) + return false; + + return true; +} + +template<typename T> bool test_classify () +{ + T x = -2.3456789; + T y = 3.6789012; + + if (std::fpclassify (x) != FP_NORMAL || std::fpclassify (y) != FP_NORMAL) + return false; + if (std::fpclassify ((T) INFINITY) != FP_INFINITE + || std::fpclassify ((T) -INFINITY) != FP_INFINITE) + return false; + if (std::fpclassify ((T) 0.0) != FP_ZERO) + return false; + if (std::fpclassify ((T) NAN) != FP_NAN) + return false; + if (!std::isfinite (x) || !std::isfinite (y)) + return false; + if (std::isfinite ((T) INFINITY) || std::isfinite ((T) -INFINITY)) + return false; + if (std::isinf (x) || std::isinf (y)) + return false; + if (!std::isinf ((T) INFINITY) || !std::isinf ((T) -INFINITY)) + return false; + if (std::isnan (x) || std::isnan (y)) + return false; + if (!std::isnan ((T) 0.0 / (T) 0.0)) + return false; + if (std::isnan (x) || std::isnan (y)) + return false; + if (!std::isnormal (x) || !std::isnormal (y)) + return false; + if (std::isnormal ((T) 0.0) || std::isnormal ((T) INFINITY) || std::isnormal ((T) NAN)) + return false; + if (!std::signbit (x) || std::signbit (y)) + return false; + + return true; +} + +template<typename T> bool test_compare () +{ + T x = 5.6789012; + T y = 8.9012345; + + if (std::isgreater (x, y)) + return false; + if (std::isgreater (x, x)) + return false; + if (std::isgreaterequal (x, y)) + return false; + if (!std::isgreaterequal (x, x)) + return false; + if (!std::isless (x, y)) + return false; + if (std::isless (x, x)) + return false; + if (!std::islessequal (x, y)) + return false; + if (!std::islessequal (x, x)) + return false; + if (!std::islessgreater (x, y)) + return false; + if (std::islessgreater (x, x)) + return false; + if (std::isunordered (x, y)) + return false; + if (!std::isunordered (x, NAN)) + return false; + return true; +} +#pragma omp end declare target + +#define RUN_TEST(func) \ +{ \ + pass++; \ + bool ok = test_##func<float> (); \ + if (!ok) { result = pass; break; } \ + pass++; \ + ok = test_##func<double> (); \ + if (!ok) { result = pass; break; } \ +} + +int main (void) +{ + int result = 0; + + #pragma omp target map (tofrom: result) + do { + int pass = 0; + + RUN_TEST (basic); + RUN_TEST (exp); + RUN_TEST (power); + RUN_TEST (trig); + RUN_TEST (hyperbolic); + RUN_TEST (erf); + RUN_TEST (gamma); + RUN_TEST (rounding); + RUN_TEST (fpmanip); + RUN_TEST (classify); + RUN_TEST (compare); + } while (false); + + return result; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__complex.C b/libgomp/testsuite/libgomp.c++/target-std__complex.C new file mode 100644 index 0000000..e392d17 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__complex.C @@ -0,0 +1,175 @@ +// { dg-do run } +// { dg-additional-options "-std=c++20" } + +#include <cmath> +#include <complex> +#include <numbers> + +using namespace std::complex_literals; + +#define FP_EQUAL(x,y) (std::abs ((x) - (y)) < 1E-6) +#define COMPLEX_EQUAL(x,y) (FP_EQUAL ((x).real (), (y).real ()) \ + && FP_EQUAL ((x).imag (), (y).imag ())) + +#pragma omp declare target +template<typename T> bool test_complex () +{ + std::complex<T> z (-1.334, 5.763); + + if (!FP_EQUAL (z.real (), (T) -1.334)) + return false; + if (!FP_EQUAL (z.imag (), (T) 5.763)) + return false; + if (!FP_EQUAL (std::abs (z), + std::sqrt (z.real () * z.real () + z.imag () * z.imag ()))) + return false; + if (!FP_EQUAL (std::arg (z), std::atan2 (z.imag (), z.real ()))) + return false; + if (!FP_EQUAL (std::norm (z), z.real () * z.real () + z.imag () * z.imag ())) + return false; + + auto conj = std::conj (z); + if (!FP_EQUAL (conj.real (), z.real ()) + || !FP_EQUAL (conj.imag (), -z.imag ())) + return false; + + if (std::proj (z) != z) + return false; + + auto infz1 = std::proj (std::complex<float> (INFINITY, -1)); + if (infz1.real () != INFINITY || infz1.imag () != (T) -0.0) + return false; + auto infz2 = std::proj (std::complex<float> (0, -INFINITY)); + if (infz2.real () != INFINITY || infz2.imag () != (T) -0.0) + return false; + + auto polarz = std::polar ((T) 1.5, std::numbers::pi_v<T> / 4); + if (!FP_EQUAL (polarz.real (), (T) 1.5 * std::cos (std::numbers::pi_v<T> / 4)) + || !FP_EQUAL (polarz.imag (), + (T) 1.5* std::sin (std::numbers::pi_v<T> / 4))) + return false; + + return true; +} + +template<typename T> bool test_complex_exp_log () +{ + std::complex<T> z (-1.724, -3.763); + + // Euler's identity + auto eulerz = std::exp (std::complex<T> (0, std::numbers::pi)); + eulerz += 1.0; + if (!COMPLEX_EQUAL (eulerz, std::complex<T> ())) + return false; + + auto my_exp_z + = std::complex<T> (std::exp (z.real ()) * std::cos (z.imag ()), + std::exp (z.real ()) * std::sin (z.imag ())); + if (!COMPLEX_EQUAL (std::exp (z), my_exp_z)) + return false; + + if (!COMPLEX_EQUAL (std::log10 (z), + std::log (z) / std::log (std::complex<T> (10)))) + return false; + + return true; +} + +template<typename T> bool test_complex_trig () +{ + std::complex<T> z (std::numbers::pi / 8, std::numbers::pi / 10); + const std::complex<T> i (0, 1); + + auto my_sin_z + = std::complex<T> (std::sin (z.real ()) * std::cosh (z.imag ()), + std::cos (z.real ()) * std::sinh (z.imag ())); + if (!COMPLEX_EQUAL (std::sin (z), my_sin_z)) + return false; + + auto my_cos_z + = std::complex<T> (std::cos (z.real ()) * std::cosh (z.imag ()), + -std::sin (z.real ()) * std::sinh (z.imag ())); + if (!COMPLEX_EQUAL (std::cos (z), my_cos_z)) + return false; + + auto my_tan_z + = std::complex<T> (std::sin (2*z.real ()), std::sinh (2*z.imag ())) + / (std::cos (2*z.real ()) + std::cosh (2*z.imag ())); + if (!COMPLEX_EQUAL (std::tan (z), my_tan_z)) + return false; + + auto my_sinh_z + = std::complex<T> (std::sinh (z.real ()) * std::cos (z.imag ()), + std::cosh (z.real ()) * std::sin (z.imag ())); + if (!COMPLEX_EQUAL (std::sinh (z), my_sinh_z)) + return false; + + auto my_cosh_z + = std::complex<T> (std::cosh (z.real ()) * std::cos (z.imag ()), + std::sinh (z.real ()) * std::sin (z.imag ())); + if (!COMPLEX_EQUAL (std::cosh (z), my_cosh_z)) + return false; + + auto my_tanh_z + = std::complex<T> (std::sinh (2*z.real ()), + std::sin (2*z.imag ())) + / (std::cosh (2*z.real ()) + std::cos (2*z.imag ())); + if (!COMPLEX_EQUAL (std::tanh (z), my_tanh_z)) + return false; + + auto my_asin_z = -i * std::log (i * z + std::sqrt ((T) 1.0 - z*z)); + if (!COMPLEX_EQUAL (std::asin (z), my_asin_z)) + return false; + + auto my_acos_z + = std::complex<T> (std::numbers::pi / 2) + + i * std::log (i * z + std::sqrt ((T) 1.0 - z*z)); + if (!COMPLEX_EQUAL (std::acos (z), my_acos_z)) + return false; + + auto my_atan_z = std::complex<T> (0, -0.5) * (std::log ((i - z) / (i + z))); + if (!COMPLEX_EQUAL (std::atan (z), my_atan_z)) + return false; + + auto my_asinh_z = std::log (z + std::sqrt (z*z + (T) 1.0)); + if (!COMPLEX_EQUAL (std::asinh (z), my_asinh_z)) + return false; + + auto my_acosh_z = std::log (z + std::sqrt (z*z - (T) 1.0)); + if (!COMPLEX_EQUAL (std::acosh (z), my_acosh_z)) + return false; + + auto my_atanh_z + = std::complex<T> (0.5) * (std::log ((T) 1.0 + z) - std::log ((T) 1.0 - z)); + if (!COMPLEX_EQUAL (std::atanh (z), my_atanh_z)) + return false; + + return true; +} +#pragma omp end declare target + +#define RUN_TEST(func) \ +{ \ + pass++; \ + bool ok = test_##func<float> (); \ + if (!ok) { result = pass; break; } \ + pass++; \ + ok = test_##func<double> (); \ + if (!ok) { result = pass; break; } \ +} + +int main (void) +{ + int result = 0; + + #pragma omp target map (tofrom: result) + do { + int pass = 0; + + RUN_TEST (complex); + RUN_TEST (complex_exp_log); + RUN_TEST (complex_trig); + } while (false); + + return result; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__deque-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__deque-concurrent-usm.C new file mode 100644 index 0000000..863a1de --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__deque-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__deque-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__deque-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__deque-concurrent.C new file mode 100644 index 0000000..9c2d6fa --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__deque-concurrent.C @@ -0,0 +1,64 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <deque> +#include <algorithm> + +#define N 50000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand (); +} + +#pragma omp declare target +bool validate (const std::deque<int> &_deque, int data[]) +{ + for (int i = 0; i < N; ++i) + if (_deque[i] != data[i] * data[i]) + return false; + return true; +} +#pragma omp end declare target + +int main (void) +{ + int data[N]; + bool ok; + + srand (time (NULL)); + init (data); + +#ifdef MEM_SHARED + std::deque<int> _deque (std::begin (data), std::end (data)); +#else + std::deque<int> _deque; +#endif + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: _deque) +#endif + { +#ifndef MEM_SHARED + #pragma omp target + new (&_deque) std::deque<int> (std::begin (data), std::end (data)); +#endif + + #pragma omp target teams distribute parallel for + for (int i = 0; i < N; ++i) + _deque[i] *= _deque[i]; + + #pragma omp target map (from: ok) + { + ok = validate (_deque, data); +#ifndef MEM_SHARED + _deque.~deque (); +#endif + } + } + + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__flat_map-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__flat_map-concurrent.C new file mode 100644 index 0000000..9e59907 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__flat_map-concurrent.C @@ -0,0 +1,71 @@ +// { dg-do run } +// { dg-additional-options "-std=c++23" } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +/* { dg-ice {TODO PR120450} { offload_target_amdgcn && { ! offload_device_shared_as } } } + { dg-excess-errors {'mkoffload' failure etc.} { xfail { offload_target_amdgcn && { ! offload_device_shared_as } } } } + (For effective-target 'offload_device_shared_as', we've got '-DMEM_SHARED', and therefore don't invoke the constructor with placement new.) */ + +#include <stdlib.h> +#include <time.h> +#include <set> +#include <flat_map> + +#define N 3000 + +void init (int data[], bool unique) +{ + std::set<int> _set; + for (int i = 0; i < N; ++i) + { + // Avoid duplicates in data array if unique is true. + do + data[i] = rand (); + while (unique && _set.count (data[i]) > 0); + _set.insert (data[i]); + } +} + +bool validate (long long sum, int keys[], int data[]) +{ + long long total = 0; + for (int i = 0; i < N; ++i) + total += (long long) keys[i] * data[i]; + return sum == total; +} + +int main (void) +{ + int keys[N], data[N]; + std::flat_map<int,int> _map; + + srand (time (NULL)); + init (keys, true); + init (data, false); + + #pragma omp target enter data map (to: keys[:N], data[:N]) map (alloc: _map) + + #pragma omp target + { +#ifndef MEM_SHARED + new (&_map) std::flat_map<int,int> (); +#endif + for (int i = 0; i < N; ++i) + _map[keys[i]] = data[i]; + } + + long long sum = 0; + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < N; ++i) + sum += (long long) keys[i] * _map[keys[i]]; + +#ifndef MEM_SHARED + #pragma omp target + _map.~flat_map (); +#endif + + #pragma omp target exit data map (release: _map) + + bool ok = validate (sum, keys, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__flat_multimap-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__flat_multimap-concurrent.C new file mode 100644 index 0000000..1dc60c8 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__flat_multimap-concurrent.C @@ -0,0 +1,70 @@ +// { dg-do run } +// { dg-additional-options "-std=c++23" } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +/* { dg-ice {TODO PR120450} { offload_target_amdgcn && { ! offload_device_shared_as } } } + { dg-excess-errors {'mkoffload' failure etc.} { xfail { offload_target_amdgcn && { ! offload_device_shared_as } } } } + (For effective-target 'offload_device_shared_as', we've got '-DMEM_SHARED', and therefore don't invoke the constructor with placement new.) */ + +#include <stdlib.h> +#include <time.h> +#include <flat_map> + +// Make sure that KEY_MAX is less than N to ensure some duplicate keys. +#define N 3000 +#define KEY_MAX 1000 + +void init (int data[], int max) +{ + for (int i = 0; i < N; ++i) + data[i] = i % max; +} + +bool validate (long long sum, int keys[], int data[]) +{ + long long total = 0; + for (int i = 0; i < N; ++i) + total += (long long) keys[i] * data[i]; + return sum == total; +} + +int main (void) +{ + int keys[N], data[N]; + std::flat_multimap<int,int> _map; + + srand (time (NULL)); + init (keys, KEY_MAX); + init (data, RAND_MAX); + + #pragma omp target enter data map (to: keys[:N], data[:N]) map (alloc: _map) + + #pragma omp target + { +#ifndef MEM_SHARED + new (&_map) std::flat_multimap<int,int> (); +#endif + for (int i = 0; i < N; ++i) + _map.insert({keys[i], data[i]}); + } + + long long sum = 0; + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < KEY_MAX; ++i) + { + auto range = _map.equal_range (i); + for (auto it = range.first; it != range.second; ++it) { + sum += (long long) it->first * it->second; + } + } + +#ifndef MEM_SHARED + #pragma omp target + _map.~flat_multimap (); +#endif + + #pragma omp target exit data map (release: _map) + + bool ok = validate (sum, keys, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__flat_multiset-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__flat_multiset-concurrent.C new file mode 100644 index 0000000..59b59bf --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__flat_multiset-concurrent.C @@ -0,0 +1,60 @@ +// { dg-do run } +// { dg-additional-options "-std=c++23" } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <flat_set> +#include <algorithm> + +// MAX should be less than N to ensure that some duplicates occur. +#define N 4000 +#define MAX 1000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand () % MAX; +} + +bool validate (int sum, int data[]) +{ + int total = 0; + for (int i = 0; i < N; ++i) + total += data[i]; + return sum == total; +} + +int main (void) +{ + int data[N]; + std::flat_multiset<int> set; + int sum = 0; + + srand (time (NULL)); + init (data); + + #pragma omp target data map (to: data[:N]) map (alloc: set) + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&set) std::flat_multiset<int> (); +#endif + for (int i = 0; i < N; ++i) + set.insert (data[i]); + } + + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < MAX; ++i) + sum += i * set.count (i); + +#ifndef MEM_SHARED + #pragma omp target + set.~flat_multiset (); +#endif + } + + bool ok = validate (sum, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__flat_set-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__flat_set-concurrent.C new file mode 100644 index 0000000..b255cd5 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__flat_set-concurrent.C @@ -0,0 +1,67 @@ +// { dg-do run } +// { dg-additional-options "-std=c++23" } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <flat_set> +#include <algorithm> + +#define N 4000 +#define MAX 16384 + +void init (int data[]) +{ + std::flat_set<int> _set; + for (int i = 0; i < N; ++i) + { + // Avoid duplicates in data array. + do + data[i] = rand () % MAX; + while (_set.count (data[i]) != 0); + _set.insert (data[i]); + } +} + +bool validate (int sum, int data[]) +{ + int total = 0; + for (int i = 0; i < N; ++i) + total += data[i]; + return sum == total; +} + +int main (void) +{ + int data[N]; + std::flat_set<int> _set; + int sum = 0; + + srand (time (NULL)); + init (data); + + #pragma omp target data map (to: data[:N]) map (alloc: _set) + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&_set) std::flat_set<int> (); +#endif + for (int i = 0; i < N; ++i) + _set.insert (data[i]); + } + + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < MAX; ++i) + if (_set.count (i) > 0) + sum += i; + +#ifndef MEM_SHARED + #pragma omp target + _set.~flat_set (); +#endif + } + + bool ok = validate (sum, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__forward_list-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__forward_list-concurrent-usm.C new file mode 100644 index 0000000..60d5cee --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__forward_list-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__forward_list-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__forward_list-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__forward_list-concurrent.C new file mode 100644 index 0000000..6b0ee65 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__forward_list-concurrent.C @@ -0,0 +1,83 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <omp.h> +#include <forward_list> +#include <algorithm> + +#define N 3000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand (); +} + +#pragma omp declare target +bool validate (const std::forward_list<int> &list, int data[]) +{ + int i = 0; + for (auto &v : list) + { + if (v != data[i] * data[i]) + return false; + ++i; + } + return true; +} +#pragma omp end declare target + +int main (void) +{ + int data[N]; + bool ok; + + srand (time (NULL)); + init (data); + +#ifdef MEM_SHARED + std::forward_list<int> list (std::begin (data), std::end (data)); +#else + std::forward_list<int> list; +#endif + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: list) +#endif + { +#ifndef MEM_SHARED + #pragma omp target + new (&list) std::forward_list<int> (std::begin (data), std::end (data)); +#endif + + #pragma omp target teams + do + { + int len = N / omp_get_num_teams () + (N % omp_get_num_teams () > 0); + int start = len * omp_get_team_num (); + if (start >= N) + break; + if (start + len >= N) + len = N - start; + auto it = list.begin (); + std::advance (it, start); + for (int i = 0; i < len; ++i) + { + *it *= *it; + ++it; + } + } while (false); + + #pragma omp target map (from: ok) + { + ok = validate (list, data); +#ifndef MEM_SHARED + list.~forward_list (); +#endif + } + } + + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__list-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__list-concurrent-usm.C new file mode 100644 index 0000000..5057bf9 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__list-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__list-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__list-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__list-concurrent.C new file mode 100644 index 0000000..1f44a17 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__list-concurrent.C @@ -0,0 +1,83 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <omp.h> +#include <list> +#include <algorithm> + +#define N 3000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand (); +} + +#pragma omp declare target +bool validate (const std::list<int> &_list, int data[]) +{ + int i = 0; + for (auto &v : _list) + { + if (v != data[i] * data[i]) + return false; + ++i; + } + return true; +} +#pragma omp end declare target + +int main (void) +{ + int data[N]; + bool ok; + + srand (time (NULL)); + init (data); + +#ifdef MEM_SHARED + std::list<int> _list (std::begin (data), std::end (data)); +#else + std::list<int> _list; +#endif + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: _list) +#endif + { +#ifndef MEM_SHARED + #pragma omp target + new (&_list) std::list<int> (std::begin (data), std::end (data)); +#endif + + #pragma omp target teams + do + { + int len = N / omp_get_num_teams () + (N % omp_get_num_teams () > 0); + int start = len * omp_get_team_num (); + if (start >= N) + break; + if (start + len >= N) + len = N - start; + auto it = _list.begin (); + std::advance (it, start); + for (int i = 0; i < len; ++i) + { + *it *= *it; + ++it; + } + } while (false); + + #pragma omp target map (from: ok) + { + ok = validate (_list, data); +#ifndef MEM_SHARED + _list.~list (); +#endif + } + } + + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__map-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__map-concurrent-usm.C new file mode 100644 index 0000000..fe37426 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__map-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__map-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__map-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__map-concurrent.C new file mode 100644 index 0000000..36556ef --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__map-concurrent.C @@ -0,0 +1,70 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <set> +#include <map> + +#define N 3000 + +void init (int data[], bool unique) +{ + std::set<int> _set; + for (int i = 0; i < N; ++i) + { + // Avoid duplicates in data array if unique is true. + do + data[i] = rand (); + while (unique && _set.find (data[i]) != _set.end ()); + _set.insert (data[i]); + } +} + +bool validate (long long sum, int keys[], int data[]) +{ + long long total = 0; + for (int i = 0; i < N; ++i) + total += (long long) keys[i] * data[i]; + return sum == total; +} + +int main (void) +{ + int keys[N], data[N]; + std::map<int,int> _map; + + srand (time (NULL)); + init (keys, true); + init (data, false); + +#ifndef MEM_SHARED + #pragma omp target enter data map (to: keys[:N], data[:N]) map (alloc: _map) +#endif + + #pragma omp target + { +#ifndef MEM_SHARED + new (&_map) std::map<int,int> (); +#endif + for (int i = 0; i < N; ++i) + _map[keys[i]] = data[i]; + } + + long long sum = 0; + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < N; ++i) + sum += (long long) keys[i] * _map[keys[i]]; + +#ifndef MEM_SHARED + #pragma omp target + _map.~map (); +#endif + +#ifndef MEM_SHARED + #pragma omp target exit data map (release: _map) +#endif + + bool ok = validate (sum, keys, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent-usm.C new file mode 100644 index 0000000..79f9245 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__multimap-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent.C new file mode 100644 index 0000000..6a4a4e8 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent.C @@ -0,0 +1,68 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <map> + +// Make sure that KEY_MAX is less than N to ensure some duplicate keys. +#define N 3000 +#define KEY_MAX 1000 + +void init (int data[], int max) +{ + for (int i = 0; i < N; ++i) + data[i] = rand () % max; +} + +bool validate (long long sum, int keys[], int data[]) +{ + long long total = 0; + for (int i = 0; i < N; ++i) + total += (long long) keys[i] * data[i]; + return sum == total; +} + +int main (void) +{ + int keys[N], data[N]; + std::multimap<int,int> _map; + + srand (time (NULL)); + init (keys, KEY_MAX); + init (data, RAND_MAX); + +#ifndef MEM_SHARED + #pragma omp target enter data map (to: keys[:N], data[:N]) map (alloc: _map) +#endif + + #pragma omp target + { +#ifndef MEM_SHARED + new (&_map) std::multimap<int,int> (); +#endif + for (int i = 0; i < N; ++i) + _map.insert({keys[i], data[i]}); + } + + long long sum = 0; + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < KEY_MAX; ++i) + { + auto range = _map.equal_range (i); + for (auto it = range.first; it != range.second; ++it) + sum += (long long) it->first * it->second; + } + +#ifndef MEM_SHARED + #pragma omp target + _map.~multimap (); +#endif + +#ifndef MEM_SHARED + #pragma omp target exit data map (release: _map) +#endif + + bool ok = validate (sum, keys, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__multiset-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__multiset-concurrent-usm.C new file mode 100644 index 0000000..2d80756 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__multiset-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__multiset-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__multiset-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__multiset-concurrent.C new file mode 100644 index 0000000..b12402e --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__multiset-concurrent.C @@ -0,0 +1,62 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <stdio.h> +#include <time.h> +#include <set> +#include <algorithm> + +// MAX should be less than N to ensure that some duplicates occur. +#define N 4000 +#define MAX 1000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand () % MAX; +} + +bool validate (int sum, int data[]) +{ + int total = 0; + for (int i = 0; i < N; ++i) + total += data[i]; + return sum == total; +} + +int main (void) +{ + int data[N]; + std::multiset<int> set; + int sum = 0; + + srand (time (NULL)); + init (data); + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: set) +#endif + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&set) std::multiset<int> (); +#endif + for (int i = 0; i < N; ++i) + set.insert (data[i]); + } + + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < MAX; ++i) + sum += i * set.count (i); + +#ifndef MEM_SHARED + #pragma omp target + set.~multiset (); +#endif + } + + bool ok = validate (sum, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__numbers.C b/libgomp/testsuite/libgomp.c++/target-std__numbers.C new file mode 100644 index 0000000..a6b3665 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__numbers.C @@ -0,0 +1,93 @@ +// { dg-do run } +// { dg-additional-options "-std=c++20" } + +#include <cmath> +#include <numbers> + +#define FP_EQUAL(x,y) (std::abs ((x) - (y)) < 1E-6) + +#pragma omp declare target +template<typename T> bool test_pi () +{ + if (!FP_EQUAL (std::sin (std::numbers::pi_v<T>), (T) 0.0)) + return false; + if (!FP_EQUAL (std::cos (std::numbers::pi_v<T>), (T) -1.0)) + return false; + if (!FP_EQUAL (std::numbers::pi_v<T> * std::numbers::inv_pi_v<T>, (T) 1.0)) + return false; + if (!FP_EQUAL (std::numbers::pi_v<T> * std::numbers::inv_sqrtpi_v<T> + * std::numbers::inv_sqrtpi_v<T>, (T) 1.0)) + return false; + return true; +} + +template<typename T> bool test_sqrt () +{ + if (!FP_EQUAL (std::numbers::sqrt2_v<T> * std::numbers::sqrt2_v<T>, (T) 2.0)) + return false; + if (!FP_EQUAL (std::numbers::sqrt3_v<T> * std::numbers::sqrt3_v<T>, (T) 3.0)) + return false; + return true; +} + +template<typename T> bool test_phi () +{ + T myphi = ((T) 1.0 + std::sqrt ((T) 5.0)) / (T) 2.0; + if (!FP_EQUAL (myphi, std::numbers::phi_v<T>)) + return false; + return true; +} + +template<typename T> bool test_log () +{ + if (!FP_EQUAL (std::log ((T) 2.0), std::numbers::ln2_v<T>)) + return false; + if (!FP_EQUAL (std::log ((T) 10.0), std::numbers::ln10_v<T>)) + return false; + if (!FP_EQUAL (std::log2 ((T) std::numbers::e), std::numbers::log2e_v<T>)) + return false; + if (!FP_EQUAL (std::log10 ((T) std::numbers::e), std::numbers::log10e_v<T>)) + return false; + return true; +} + +template<typename T> bool test_egamma () +{ + T myegamma = 0.0; + #pragma omp parallel for reduction(+:myegamma) + for (int k = 2; k < 100000; ++k) + myegamma += (std::riemann_zeta (k) - 1) / k; + myegamma = (T) 1 - myegamma; + if (!FP_EQUAL (myegamma, std::numbers::egamma_v<T>)) + return false; + return true; +} +#pragma omp end declare target + +#define RUN_TEST(func) \ +{ \ + pass++; \ + bool ok = test_##func<float> (); \ + if (!ok) { result = pass; break; } \ + pass++; \ + ok = test_##func<double> (); \ + if (!ok) { result = pass; break; } \ +} + +int main (void) +{ + int result = 0; + + #pragma omp target map (tofrom: result) + do { + int pass = 0; + + RUN_TEST (pi); + RUN_TEST (sqrt); + RUN_TEST (phi); + RUN_TEST (log); + RUN_TEST (egamma); + } while (false); + + return result; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__set-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__set-concurrent-usm.C new file mode 100644 index 0000000..54f62e3 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__set-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__set-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__set-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__set-concurrent.C new file mode 100644 index 0000000..cd23128 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__set-concurrent.C @@ -0,0 +1,68 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <set> +#include <algorithm> + +#define N 4000 +#define MAX 16384 + +void init (int data[]) +{ + std::set<int> _set; + for (int i = 0; i < N; ++i) + { + // Avoid duplicates in data array. + do + data[i] = rand () % MAX; + while (_set.find (data[i]) != _set.end ()); + _set.insert (data[i]); + } +} + +bool validate (int sum, int data[]) +{ + int total = 0; + for (int i = 0; i < N; ++i) + total += data[i]; + return sum == total; +} + +int main (void) +{ + int data[N]; + std::set<int> _set; + int sum = 0; + + srand (time (NULL)); + init (data); + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: _set) +#endif + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&_set) std::set<int> (); +#endif + for (int i = 0; i < N; ++i) + _set.insert (data[i]); + } + + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < MAX; ++i) + if (_set.find (i) != _set.end ()) + sum += i; + +#ifndef MEM_SHARED + #pragma omp target + _set.~set (); +#endif + } + + bool ok = validate (sum, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__span-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__span-concurrent-usm.C new file mode 100644 index 0000000..7ef16bf --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__span-concurrent-usm.C @@ -0,0 +1,7 @@ +// { dg-additional-options "-std=c++20" } + +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__span-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__span-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__span-concurrent.C new file mode 100644 index 0000000..046b3c1 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__span-concurrent.C @@ -0,0 +1,66 @@ +// { dg-do run } +// { dg-additional-options "-std=c++20" } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <span> + +#define N 64 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand (); +} + +#pragma omp declare target +bool validate (const std::span<int, N> &span, int data[]) +{ + for (int i = 0; i < N; ++i) + if (span[i] != data[i] * data[i]) + return false; + return true; +} +#pragma omp end declare target + +int main (void) +{ + int data[N]; + bool ok; + int elements[N]; + std::span<int, N> span(elements); + + srand (time (NULL)); + init (data); + +#ifndef MEM_SHARED + #pragma omp target enter data map (to: data[:N]) map (alloc: elements, span) +#endif + + #pragma omp target + { +#ifndef MEM_SHARED + new (&span) std::span<int, N> (elements); +#endif + std::copy (data, data + N, span.begin ()); + } + + #pragma omp target teams distribute parallel for + for (int i = 0; i < N; ++i) + span[i] *= span[i]; + + #pragma omp target map (from: ok) + { + ok = validate (span, data); +#ifndef MEM_SHARED + span.~span (); +#endif + } + +#ifndef MEM_SHARED + #pragma omp target exit data map (release: elements, span) +#endif + + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__unordered_map-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__unordered_map-concurrent.C new file mode 100644 index 0000000..00d7943 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__unordered_map-concurrent.C @@ -0,0 +1,66 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <set> +#include <unordered_map> + +#define N 3000 + +void init (int data[], bool unique) +{ + std::set<int> _set; + for (int i = 0; i < N; ++i) + { + // Avoid duplicates in data array if unique is true. + do + data[i] = rand (); + while (unique && _set.count (data[i]) > 0); + _set.insert (data[i]); + } +} + +bool validate (long long sum, int keys[], int data[]) +{ + long long total = 0; + for (int i = 0; i < N; ++i) + total += (long long) keys[i] * data[i]; + return sum == total; +} + +int main (void) +{ + int keys[N], data[N]; + std::unordered_map<int,int> _map; + + srand (time (NULL)); + init (keys, true); + init (data, false); + + #pragma omp target enter data map (to: keys[:N], data[:N]) map (alloc: _map) + + #pragma omp target + { +#ifndef MEM_SHARED + new (&_map) std::unordered_map<int,int> (); +#endif + for (int i = 0; i < N; ++i) + _map[keys[i]] = data[i]; + } + + long long sum = 0; + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < N; ++i) + sum += (long long) keys[i] * _map[keys[i]]; + +#ifndef MEM_SHARED + #pragma omp target + _map.~unordered_map (); +#endif + + #pragma omp target exit data map (release: _map) + + bool ok = validate (sum, keys, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__unordered_multimap-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__unordered_multimap-concurrent.C new file mode 100644 index 0000000..2567634 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__unordered_multimap-concurrent.C @@ -0,0 +1,65 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <unordered_map> + +// Make sure that KEY_MAX is less than N to ensure some duplicate keys. +#define N 3000 +#define KEY_MAX 1000 + +void init (int data[], int max) +{ + for (int i = 0; i < N; ++i) + data[i] = i % max; +} + +bool validate (long long sum, int keys[], int data[]) +{ + long long total = 0; + for (int i = 0; i < N; ++i) + total += (long long) keys[i] * data[i]; + return sum == total; +} + +int main (void) +{ + int keys[N], data[N]; + std::unordered_multimap<int,int> _map; + + srand (time (NULL)); + init (keys, KEY_MAX); + init (data, RAND_MAX); + + #pragma omp target enter data map (to: keys[:N], data[:N]) map (alloc: _map) + + #pragma omp target + { +#ifndef MEM_SHARED + new (&_map) std::unordered_multimap<int,int> (); +#endif + for (int i = 0; i < N; ++i) + _map.insert({keys[i], data[i]}); + } + + long long sum = 0; + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < KEY_MAX; ++i) + { + auto range = _map.equal_range (i); + for (auto it = range.first; it != range.second; ++it) { + sum += (long long) it->first * it->second; + } + } + +#ifndef MEM_SHARED + #pragma omp target + _map.~unordered_multimap (); +#endif + + #pragma omp target exit data map (release: _map) + + bool ok = validate (sum, keys, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__unordered_multiset-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__unordered_multiset-concurrent.C new file mode 100644 index 0000000..da6c875 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__unordered_multiset-concurrent.C @@ -0,0 +1,59 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <unordered_set> +#include <algorithm> + +// MAX should be less than N to ensure that some duplicates occur. +#define N 4000 +#define MAX 1000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand () % MAX; +} + +bool validate (int sum, int data[]) +{ + int total = 0; + for (int i = 0; i < N; ++i) + total += data[i]; + return sum == total; +} + +int main (void) +{ + int data[N]; + std::unordered_multiset<int> set; + int sum = 0; + + srand (time (NULL)); + init (data); + + #pragma omp target data map (to: data[:N]) map (alloc: set) + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&set) std::unordered_multiset<int> (); +#endif + for (int i = 0; i < N; ++i) + set.insert (data[i]); + } + + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < MAX; ++i) + sum += i * set.count (i); + +#ifndef MEM_SHARED + #pragma omp target + set.~unordered_multiset (); +#endif + } + + bool ok = validate (sum, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__unordered_set-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__unordered_set-concurrent.C new file mode 100644 index 0000000..b7bd935 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__unordered_set-concurrent.C @@ -0,0 +1,66 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <unordered_set> +#include <algorithm> + +#define N 4000 +#define MAX 16384 + +void init (int data[]) +{ + std::unordered_set<int> _set; + for (int i = 0; i < N; ++i) + { + // Avoid duplicates in data array. + do + data[i] = rand () % MAX; + while (_set.count (data[i]) != 0); + _set.insert (data[i]); + } +} + +bool validate (int sum, int data[]) +{ + int total = 0; + for (int i = 0; i < N; ++i) + total += data[i]; + return sum == total; +} + +int main (void) +{ + int data[N]; + std::unordered_set<int> _set; + int sum = 0; + + srand (time (NULL)); + init (data); + + #pragma omp target data map (to: data[:N]) map (alloc: _set) + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&_set) std::unordered_set<int> (); +#endif + for (int i = 0; i < N; ++i) + _set.insert (data[i]); + } + + #pragma omp target teams distribute parallel for reduction (+:sum) + for (int i = 0; i < MAX; ++i) + if (_set.count (i) > 0) + sum += i; + +#ifndef MEM_SHARED + #pragma omp target + _set.~unordered_set (); +#endif + } + + bool ok = validate (sum, data); + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__valarray-1.C b/libgomp/testsuite/libgomp.c++/target-std__valarray-1.C new file mode 100644 index 0000000..865cde2 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__valarray-1.C @@ -0,0 +1,179 @@ +// { dg-additional-options -std=c++20 } +// { dg-output-file target-std__valarray-1.output } + +#include <valarray> +#include <ostream> +#include <sstream> + + +/*TODO Work around PR118484 "ICE during IPA pass: cp, segfault in determine_versionability ipa-cp.cc:467". + +We can't: + + #pragma omp declare target(std::basic_streambuf<char, std::char_traits<char>>::basic_streambuf) + +... because: + + error: overloaded function name ‘std::basic_streambuf<char>::__ct ’ in clause ‘enter’ + +Therefore, use dummy classes in '#pragma omp declare target': +*/ + +#pragma omp declare target + +// For 'std::basic_streambuf<char, std::char_traits<char> >::basic_streambuf': + +class dummy_basic_streambuf__char + : public std::basic_streambuf<char> +{ +public: + dummy_basic_streambuf__char() {} +}; + +// For 'std::basic_ios<char, std::char_traits<char> >::basic_ios()': + +class dummy_basic_ios__char + : public std::basic_ios<char> +{ +public: + dummy_basic_ios__char() {} +}; + +#pragma omp end declare target + + +int main() +{ + // Due to PR120021 "Offloading vs. C++ 'std::initializer_list'", we can't construct these on the device. + std::initializer_list<int> v1_i = {10, 20, 30, 40, 50}; + const int *v1_i_data = std::data(v1_i); + size_t v1_i_size = v1_i.size(); + std::initializer_list<int> v2_i = {5, 4, 3, 2, 1}; + const int *v2_i_data = std::data(v2_i); + size_t v2_i_size = v2_i.size(); + std::initializer_list<int> shiftData_i = {1, 2, 3, 4, 5}; + const int *shiftData_i_data = std::data(shiftData_i); + size_t shiftData_i_size = shiftData_i.size(); +#pragma omp target \ + defaultmap(none) \ + map(to: v1_i_data[:v1_i_size], v1_i_size, \ + v2_i_data[:v2_i_size], v2_i_size, \ + shiftData_i_data[:shiftData_i_size], shiftData_i_size) + { + /* Manually set up a buffer we can stream into, similar to 'cout << [...]', and print it at the end of region. */ + std::stringbuf out_b; + std::ostream out(&out_b); + + std::valarray<int> v1(v1_i_data, v1_i_size); + out << "\nv1:"; + for (auto val : v1) + out << " " << val; + + std::valarray<int> v2(v2_i_data, v2_i_size); + out << "\nv2:"; + for (auto val : v2) + out << " " << val; + + std::valarray<int> sum = v1 + v2; + out << "\nv1 + v2:"; + for (auto val : sum) + out << " " << val; + + std::valarray<int> diff = v1 - v2; + out << "\nv1 - v2:"; + for (auto val : diff) + out << " " << val; + + std::valarray<int> product = v1 * v2; + out << "\nv1 * v2:"; + for (auto val : product) + out << " " << val; + + std::valarray<int> quotient = v1 / v2; + out << "\nv1 / v2:"; + for (auto val : quotient) + out << " " << val; + + std::valarray<int> squares = pow(v1, 2); + out << "\npow(v1, 2):"; + for (auto val : squares) + out << " " << val; + + std::valarray<int> sinhs = sinh(v2); + out << "\nsinh(v2):"; + for (auto val : sinhs) + out << " " << val; + + std::valarray<int> logs = log(v1 * v2); + out << "\nlog(v1 * v2):"; + for (auto val : logs) + out << " " << val; + + std::valarray<int> data(12); + for (size_t i = 0; i < data.size(); ++i) + data[i] = i; + out << "\nOriginal array:"; + for (auto val : data) + out << " " << val; + + std::slice slice1(2, 5, 1); + std::valarray<int> sliced1 = data[slice1]; + out << "\nSlice(2, 5, 1):"; + for (auto val : sliced1) + out << " " << val; + + std::slice slice2(1, 4, 3); + std::valarray<int> sliced2 = data[slice2]; + out << "\nSlice(1, 4, 3):"; + for (auto val : sliced2) + out << " " << val; + + data[slice1] = 99; + out << "\nArray after slice modification:"; + for (auto val : data) + out << " " << val; + + std::valarray<bool> mask = (v1 > 20); + out << "\nElements of v1 > 20:"; + for (size_t i = 0; i < v1.size(); ++i) + { + if (mask[i]) + out << " " << v1[i]; + } + + std::valarray<int> masked = v1[mask]; + out << "\nMasked array:"; + for (auto val : masked) + out << " " << val; + + std::valarray<int> shiftData(shiftData_i_data, shiftData_i_size); + out << "\nOriginal shiftData:"; + for (auto val : shiftData) + out << " " << val; + + std::valarray<int> shifted = shiftData.shift(2); + out << "\nshift(2):"; + for (auto val : shifted) + out << " " << val; + + std::valarray<int> cshifted = shiftData.cshift(-1); + out << "\ncshift(-1):"; + for (auto val : cshifted) + out << " " << val; + + out << "\nSum(v1): " << v1.sum(); + out << "\nMin(v1): " << v1.min(); + out << "\nMax(v1): " << v1.max(); + + out << "\n"; + + /* Terminate with a NUL. Otherwise, we'd have to use: + __builtin_printf("%.*s", (int) out_b_sv.size(), out_b_sv.data()); + ... which nvptx 'printf', as implemented via PTX 'vprintf', doesn't support (TODO). */ + out << '\0'; + std::string_view out_b_sv = out_b.view(); + __builtin_printf("%s", out_b_sv.data()); + } + + return 0; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__valarray-1.output b/libgomp/testsuite/libgomp.c++/target-std__valarray-1.output new file mode 100644 index 0000000..c441e06 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__valarray-1.output @@ -0,0 +1,22 @@ + +v1: 10 20 30 40 50 +v2: 5 4 3 2 1 +v1 + v2: 15 24 33 42 51 +v1 - v2: 5 16 27 38 49 +v1 * v2: 50 80 90 80 50 +v1 / v2: 2 5 10 20 50 +pow(v1, 2): 100 400 900 1600 2500 +sinh(v2): 74 27 10 3 1 +log(v1 * v2): 3 4 4 4 3 +Original array: 0 1 2 3 4 5 6 7 8 9 10 11 +Slice(2, 5, 1): 2 3 4 5 6 +Slice(1, 4, 3): 1 4 7 10 +Array after slice modification: 0 1 99 99 99 99 99 7 8 9 10 11 +Elements of v1 > 20: 30 40 50 +Masked array: 30 40 50 +Original shiftData: 1 2 3 4 5 +shift(2): 3 4 5 0 0 +cshift(-1): 5 1 2 3 4 +Sum(v1): 150 +Min(v1): 10 +Max(v1): 50 diff --git a/libgomp/testsuite/libgomp.c++/target-std__valarray-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__valarray-concurrent-usm.C new file mode 100644 index 0000000..41ec80e --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__valarray-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__valarray-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__valarray-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__valarray-concurrent.C new file mode 100644 index 0000000..8933072b --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__valarray-concurrent.C @@ -0,0 +1,66 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <valarray> + +#define N 50000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand (); +} + +#pragma omp declare target +bool validate (const std::valarray<int> &arr, int data[]) +{ + for (int i = 0; i < N; ++i) + if (arr[i] != data[i] * data[i] + i) + return false; + return true; +} +#pragma omp end declare target + +int main (void) +{ + int data[N]; + bool ok; + + srand (time (NULL)); + init (data); + +#ifdef MEM_SHARED + std::valarray<int> arr (data, N); +#else + std::valarray<int> arr; +#endif + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: arr) +#endif + { + #pragma omp target + { +#ifndef MEM_SHARED + new (&arr) std::valarray<int> (data, N); +#endif + arr *= arr; + } + + #pragma omp target teams distribute parallel for + for (int i = 0; i < N; ++i) + arr[i] += i; + + #pragma omp target map (from: ok) + { + ok = validate (arr, data); +#ifndef MEM_SHARED + arr.~valarray (); +#endif + } + } + + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c++/target-std__vector-concurrent-usm.C b/libgomp/testsuite/libgomp.c++/target-std__vector-concurrent-usm.C new file mode 100644 index 0000000..967bff3 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__vector-concurrent-usm.C @@ -0,0 +1,5 @@ +#pragma omp requires unified_shared_memory self_maps + +#define MEM_SHARED + +#include "target-std__vector-concurrent.C" diff --git a/libgomp/testsuite/libgomp.c++/target-std__vector-concurrent.C b/libgomp/testsuite/libgomp.c++/target-std__vector-concurrent.C new file mode 100644 index 0000000..a94b4cf --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-std__vector-concurrent.C @@ -0,0 +1,63 @@ +// { dg-do run } +// { dg-additional-options -DMEM_SHARED { target offload_device_shared_as } } + +#include <stdlib.h> +#include <time.h> +#include <vector> + +#define N 50000 + +void init (int data[]) +{ + for (int i = 0; i < N; ++i) + data[i] = rand (); +} + +#pragma omp declare target +bool validate (const std::vector<int> &vec, int data[]) +{ + for (int i = 0; i < N; ++i) + if (vec[i] != data[i] * data[i]) + return false; + return true; +} +#pragma omp end declare target + +int main (void) +{ + int data[N]; + bool ok; + + srand (time (NULL)); + init (data); + +#ifdef MEM_SHARED + std::vector<int> vec (data, data + N); +#else + std::vector<int> vec; +#endif + +#ifndef MEM_SHARED + #pragma omp target data map (to: data[:N]) map (alloc: vec) +#endif + { +#ifndef MEM_SHARED + #pragma omp target + new (&vec) std::vector<int> (data, data + N); +#endif + + #pragma omp target teams distribute parallel for + for (int i = 0; i < N; ++i) + vec[i] *= vec[i]; + + #pragma omp target map (from: ok) + { + ok = validate (vec, data); +#ifndef MEM_SHARED + vec.~vector (); +#endif + } + } + + return ok ? 0 : 1; +} diff --git a/libgomp/testsuite/libgomp.c-c++-common/target-abi-struct-1-O0.c b/libgomp/testsuite/libgomp.c-c++-common/target-abi-struct-1-O0.c index 35ec75d..9bf949a 100644 --- a/libgomp/testsuite/libgomp.c-c++-common/target-abi-struct-1-O0.c +++ b/libgomp/testsuite/libgomp.c-c++-common/target-abi-struct-1-O0.c @@ -1,3 +1,3 @@ /* { dg-additional-options -O0 } */ -#include "../libgomp.oacc-c-c++-common/abi-struct-1.c" +#include "target-abi-struct-1.c" diff --git a/libgomp/testsuite/libgomp.c-c++-common/target-abi-struct-1.c b/libgomp/testsuite/libgomp.c-c++-common/target-abi-struct-1.c new file mode 100644 index 0000000..d9268af --- /dev/null +++ b/libgomp/testsuite/libgomp.c-c++-common/target-abi-struct-1.c @@ -0,0 +1 @@ +#include "../libgomp.oacc-c-c++-common/abi-struct-1.c" diff --git a/libgomp/testsuite/libgomp.fortran/metadirective-1.f90 b/libgomp/testsuite/libgomp.fortran/metadirective-1.f90 index 7b3e09f..d6f4d5b 100644 --- a/libgomp/testsuite/libgomp.fortran/metadirective-1.f90 +++ b/libgomp/testsuite/libgomp.fortran/metadirective-1.f90 @@ -1,4 +1,5 @@ -! { dg-do run } +! { dg-do run { target { ! offload_target_nvptx } } } +! { dg-do compile { target offload_target_nvptx } } program test implicit none @@ -33,6 +34,10 @@ program test contains subroutine f (x, y, z) integer :: x(N), y(N), z(N) + ! The following fails as on the host the target side cannot be + ! resolved - and the 'teams' or not status affects how 'target' + ! is called. -> See PR118694, esp. comment 9. + ! Note also the dg-do compile above for offload_target_nvptx !$omp target map (to: x, y) map(from: z) block @@ -43,6 +48,7 @@ contains z(i) = x(i) * y(i) enddo end block + ! { dg-bogus "'target' construct with nested 'teams' construct contains directives outside of the 'teams' construct" "PR118694" { xfail offload_target_nvptx } .-9 } */ end subroutine subroutine g (x, y, z) integer :: x(N), y(N), z(N) @@ -56,6 +62,7 @@ contains z(i) = x(i) * y(i) enddo end block + ! { dg-bogus "'target' construct with nested 'teams' construct contains directives outside of the 'teams' construct" "PR118694" { xfail offload_target_nvptx } .-9 } */ !$omp end target end subroutine end program diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/abi-struct-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/abi-struct-1.c index 8078655..4b54171 100644 --- a/libgomp/testsuite/libgomp.oacc-c-c++-common/abi-struct-1.c +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/abi-struct-1.c @@ -1,6 +1,10 @@ /* Inspired by 'gcc.target/nvptx/abi-struct-arg.c', 'gcc.target/nvptx/abi-struct-ret.c'. */ -/* See also '../libgomp.c-c++-common/target-abi-struct-1-O0.c'. */ +/* See also '../libgomp.c-c++-common/target-abi-struct-1.c'. */ + +/* To exercise PR119835 (if optimizations enabled): disable inlining, so that + GIMPLE passes still see the functions that return aggregate types. */ +#pragma GCC optimize "-fno-inline" typedef struct {} empty; /* See 'gcc/doc/extend.texi', "Empty Structures". */ typedef struct {char a;} schar; diff --git a/libstdc++-v3/ChangeLog b/libstdc++-v3/ChangeLog index d7cd8f4..4dfe958 100644 --- a/libstdc++-v3/ChangeLog +++ b/libstdc++-v3/ChangeLog @@ -1,3 +1,47 @@ +2025-06-01 Patrick Palka <ppalka@redhat.com> + + Backported from master: + 2025-05-29 Patrick Palka <ppalka@redhat.com> + + * include/std/flat_map (_Flat_map_impl::operator==): Compare + keys and values separately. + +2025-06-01 Patrick Palka <ppalka@redhat.com> + + Backported from master: + 2025-05-29 Patrick Palka <ppalka@redhat.com> + Jonathan Wakely <jwakely@redhat.com> + + PR libstdc++/120465 + * include/std/flat_map (_Flat_map_impl::_M_erase_if): Use a + projection with ranges::remove_if to pass a pair instead of + a tuple to the predicate. + * testsuite/23_containers/flat_map/1.cc (test07): Strengthen + to expect the argument passed to the predicate is a pair. + * testsuite/23_containers/flat_multimap/1.cc (test07): Likewise. + +2025-05-27 Jonathan Wakely <jwakely@redhat.com> + + Backported from master: + 2025-05-22 Jonathan Wakely <jwakely@redhat.com> + + * testsuite/util/pstl/test_utils.h (ForwardIterator::operator++): + Fix return type. + (BidirectionalIterator::operator++): Likewise. + (BidirectionalIterator::operator--): Likewise. + +2025-05-27 Jonathan Wakely <jwakely@redhat.com> + + Backported from master: + 2025-05-22 Jonathan Wakely <jwakely@redhat.com> + + PR libstdc++/120367 + * include/bits/stl_vector.h (_M_range_initialize): Initialize + _M_impl._M_finish. + * testsuite/23_containers/vector/cons/from_range.cc: Check with + a type that throws on construction. + exceptions during construction. + 2025-05-20 Jonathan Wakely <jwakely@redhat.com> Backported from master: diff --git a/libstdc++-v3/include/bits/stl_vector.h b/libstdc++-v3/include/bits/stl_vector.h index aff9d5d..4861edb 100644 --- a/libstdc++-v3/include/bits/stl_vector.h +++ b/libstdc++-v3/include/bits/stl_vector.h @@ -1981,8 +1981,9 @@ _GLIBCXX_BEGIN_NAMESPACE_CONTAINER _M_range_initialize_n(_Iterator __first, _Sentinel __last, size_type __n) { - pointer __start = this->_M_impl._M_start = + pointer __start = this->_M_allocate(_S_check_init_len(__n, _M_get_Tp_allocator())); + this->_M_impl._M_start = this->_M_impl._M_finish = __start; this->_M_impl._M_end_of_storage = __start + __n; this->_M_impl._M_finish = std::__uninitialized_copy_a(_GLIBCXX_MOVE(__first), __last, diff --git a/libstdc++-v3/include/std/flat_map b/libstdc++-v3/include/std/flat_map index 6593988..4bd4963 100644 --- a/libstdc++-v3/include/std/flat_map +++ b/libstdc++-v3/include/std/flat_map @@ -873,7 +873,10 @@ _GLIBCXX_BEGIN_NAMESPACE_VERSION [[nodiscard]] friend bool operator==(const _Derived& __x, const _Derived& __y) - { return std::equal(__x.begin(), __x.end(), __y.begin(), __y.end()); } + { + return __x._M_cont.keys == __y._M_cont.keys + && __x._M_cont.values == __y._M_cont.values; + } template<typename _Up = value_type> [[nodiscard]] @@ -895,7 +898,10 @@ _GLIBCXX_BEGIN_NAMESPACE_VERSION { auto __guard = _M_make_clear_guard(); auto __zv = views::zip(_M_cont.keys, _M_cont.values); - auto __sr = ranges::remove_if(__zv, __pred); + auto __sr = ranges::remove_if(__zv, __pred, + [](const auto& __e) { + return const_reference(__e); + }); auto __erased = __sr.size(); erase(end() - __erased, end()); __guard._M_disable(); diff --git a/libstdc++-v3/testsuite/23_containers/flat_map/1.cc b/libstdc++-v3/testsuite/23_containers/flat_map/1.cc index 38fd449..9d99796 100644 --- a/libstdc++-v3/testsuite/23_containers/flat_map/1.cc +++ b/libstdc++-v3/testsuite/23_containers/flat_map/1.cc @@ -245,8 +245,9 @@ void test07() { // PR libstdc++/119427 - std::erase_if(std::flat_foo) does not work + // PR libstdc++/120465 - erase_if for flat_map calls predicate with incorrect type std::flat_map<int, int> m = {std::pair{1, 2}, {3, 4}, {5, 6}}; - auto n = std::erase_if(m, [](auto x) { auto [k,v] = x; return k == 1 || v == 6; }); + auto n = std::erase_if(m, [](auto x) { return x.first == 1 || x.second == 6; }); VERIFY( n == 2 ); VERIFY( std::ranges::equal(m, (std::pair<int,int>[]){{3,4}}) ); } diff --git a/libstdc++-v3/testsuite/23_containers/flat_multimap/1.cc b/libstdc++-v3/testsuite/23_containers/flat_multimap/1.cc index 79fbc1a..0010955 100644 --- a/libstdc++-v3/testsuite/23_containers/flat_multimap/1.cc +++ b/libstdc++-v3/testsuite/23_containers/flat_multimap/1.cc @@ -223,8 +223,9 @@ void test07() { // PR libstdc++/119427 - std::erase_if(std::flat_foo) does not work + // PR libstdc++/120465 - erase_if for flat_map calls predicate with incorrect type std::flat_multimap<int, int> m = {std::pair{1, 2}, {3, 4}, {3, 3}, {5, 6}, {6, 6}}; - auto n = std::erase_if(m, [](auto x) { auto [k,v] = x; return k == 1 || v == 6; }); + auto n = std::erase_if(m, [](auto x) { return x.first == 1 || x.second == 6; }); VERIFY( n == 3 ); VERIFY( std::ranges::equal(m, (std::pair<int,int>[]){{3,4},{3,3}}) ); } diff --git a/libstdc++-v3/testsuite/23_containers/vector/cons/from_range.cc b/libstdc++-v3/testsuite/23_containers/vector/cons/from_range.cc index 7a62645..3784b9c 100644 --- a/libstdc++-v3/testsuite/23_containers/vector/cons/from_range.cc +++ b/libstdc++-v3/testsuite/23_containers/vector/cons/from_range.cc @@ -106,8 +106,30 @@ test_constexpr() return true; } +void +test_pr120367() +{ +#ifdef __cpp_exceptions + struct X + { + X(int) { throw 1; } // Cannot successfully construct an X. + ~X() { VERIFY(false); } // So should never need to destroy one. + }; + + try + { + int i[1]{}; + std::vector<X> v(std::from_range, i); + } + catch (int) + { + } +#endif +} + int main() { test_ranges(); static_assert( test_constexpr() ); + test_pr120367(); } diff --git a/libstdc++-v3/testsuite/util/pstl/test_utils.h b/libstdc++-v3/testsuite/util/pstl/test_utils.h index 55b5100..9c61a714 100644 --- a/libstdc++-v3/testsuite/util/pstl/test_utils.h +++ b/libstdc++-v3/testsuite/util/pstl/test_utils.h @@ -154,7 +154,7 @@ class ForwardIterator explicit ForwardIterator(Iterator i) : my_iterator(i) {} reference operator*() const { return *my_iterator; } Iterator operator->() const { return my_iterator; } - ForwardIterator + ForwardIterator& operator++() { ++my_iterator; @@ -194,13 +194,13 @@ class BidirectionalIterator : public ForwardIterator<Iterator, IteratorTag> explicit BidirectionalIterator(Iterator i) : base_type(i) {} BidirectionalIterator(const base_type& i) : base_type(i.iterator()) {} - BidirectionalIterator + BidirectionalIterator& operator++() { ++base_type::my_iterator; return *this; } - BidirectionalIterator + BidirectionalIterator& operator--() { --base_type::my_iterator; |