aboutsummaryrefslogtreecommitdiff
path: root/libgomp/ChangeLog
AgeCommit message (Collapse)Author
2019-10-16[arm] fix bootstrap failure due to uninitialized warningrearnsha
The Arm port is failing bootstrap because GCC is now warning about an unitialized array. The code is complex enough that I certainly can't be sure the compiler is wrong, so perhaps the best fix here is just to memset the entire array before use. * config/arm/arm.c (neon_valid_immediate): Clear bytes before use. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@277073 138bc75d-0d04-0410-961f-82ee72b054a4
2019-09-13libgomp plugin - init stringburnus
libgomp/ 2019-09-13 Tobias Burnus <tobias@codesourcery.com> * plugin/plugin-hsa.c (hsa_warn, hsa_fatal, hsa_error): Ensure string is initialized. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@275703 138bc75d-0d04-0410-961f-82ee72b054a4
2019-09-06Fix GCC_LINUX_FUTEX to work with C99 compilersfw
Without this change, libstdc++ is built without futex symbols if GCC rejects implicit function declarations by default. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@275454 138bc75d-0d04-0410-961f-82ee72b054a4
2019-09-032019-09-03 Chung-Lin Tang <cltang@codesourcery.com>cltang
libatomic/ PR other/79543 * acinclude.m4 (LIBAT_CHECK_LINKER_FEATURES): Fix GNU ld --version scanning to conform to the GNU Coding Standards. * configure: Regenerate. libffi/ PR other/79543 * acinclude.m4 (LIBAT_CHECK_LINKER_FEATURES): Fix GNU ld --version scanning to conform to the GNU Coding Standards. * configure: Regenerate. libgomp/ PR other/79543 * acinclude.m4 (LIBGOMP_CHECK_LINKER_FEATURES): Fix GNU ld --version scanning to conform to the GNU Coding Standards. * configure: Regenerate. libitm/ PR other/79543 * acinclude.m4 (LIBITM_CHECK_LINKER_FEATURES): Fix GNU ld --version scanning to conform to the GNU Coding Standards. * configure: Regenerate. libstdc++-v3/ PR other/79543 * acinclude.m4 (GLIBCXX_CHECK_LINKER_FEATURES): Fix GNU ld --version scanning to conform to the GNU Coding Standards. * configure: Regenerate. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@275341 138bc75d-0d04-0410-961f-82ee72b054a4
2019-08-28 PR libgomp/91530jakub
* config/i386/sse.md (vec_shl_<mode>, vec_shr_<mode>): Use V_128 iterator instead of VI_128. * testsuite/libgomp.c/scan-21.c: New test. * testsuite/libgomp.c/scan-22.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@274984 138bc75d-0d04-0410-961f-82ee72b054a4
2019-08-27 PR libgomp/91530jakub
* testsuite/libgomp.c/scan-11.c: Add -msse2 option for sse2_runtime targets. * testsuite/libgomp.c/scan-12.c: Likewise. * testsuite/libgomp.c/scan-13.c: Likewise. * testsuite/libgomp.c/scan-14.c: Likewise. * testsuite/libgomp.c/scan-15.c: Likewise. * testsuite/libgomp.c/scan-16.c: Likewise. * testsuite/libgomp.c/scan-17.c: Likewise. * testsuite/libgomp.c/scan-18.c: Likewise. * testsuite/libgomp.c/scan-19.c: Likewise. * testsuite/libgomp.c/scan-20.c: Likewise. * testsuite/libgomp.c++/scan-9.C: Likewise. * testsuite/libgomp.c++/scan-10.C: Likewise. * testsuite/libgomp.c++/scan-11.C: Likewise. * testsuite/libgomp.c++/scan-12.C: Likewise. * testsuite/libgomp.c++/scan-14.C: Likewise. * testsuite/libgomp.c++/scan-15.C: Likewise. * testsuite/libgomp.c++/scan-13.C: Likewise. Use sse2_runtime instead of i?86-*-* x86_64-*-* as target for scan-tree-dump-times. * testsuite/libgomp.c++/scan-16.C: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@274947 138bc75d-0d04-0410-961f-82ee72b054a4
2019-08-172019-08-17 Thomas Koenig <tkoenig@gcc.gnu.org>tkoenig
PR fortran/91473 * testsuite/libgomp.fortran/appendix-a/a.28.5.f90: Add -std=legacy so invalid code in the test case is accepted. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@274602 138bc75d-0d04-0410-961f-82ee72b054a4
2019-08-122019-08-12 Thomas Koenig <tkoenig@gcc.gnu.org>tkoenig
PR fortran/91424 * frontend-passes.c (do_subscript): Do not warn for an expression a second time. Do not warn about a zero-trip loop. (doloop_warn): Also look at contained namespaces. 2019-08-12 Thomas Koenig <tkoenig@gcc.gnu.org> PR fortran/91424 * gfortran.dg/do_subscript_3.f90: New test. * gfortran.dg/do_subscript_4.f90: New test. * gfortran.dg/pr70754.f90: Use indices that to not overflow. 2019-08-12 Thomas Koenig <tkoenig@gcc.gnu.org> PR fortran/91422 * testsuite/libgomp.oacc-fortran/routine-7.f90: Correct array dimension. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@274320 138bc75d-0d04-0410-961f-82ee72b054a4
2019-08-08 * gimplify.c (omp_add_variable): Use GOVD_PRIVATE | GOVD_EXPLICITjakub
for VLA helper variables on target data even if not GOVD_FIRSTPRIVATE. (gimplify_scan_omp_clauses): For OMP_CLAUSE_USE_DEVICE_* use just GOVD_EXPLICIT flags. (gimplify_omp_workshare): For OMP_TARGET_DATA move all OMP_CLAUSE_USE_DEVICE_* clauses to the end of clauses chain. * omp-low.c (scan_sharing_clauses): For OMP_CLAUSE_USE_DEVICE_* call install_var_field with mask 11 instead of 3. (lower_omp_target): For OMP_CLAUSE_USE_DEVICE_* use pass (splay_tree_key) &DECL_UID (var) to build_sender_ref instead of var. gcc/c/ * c-typeck.c (c_finish_omp_clauses): For C_ORT_OMP OMP_CLAUSE_USE_DEVICE_* clauses use oacc_reduction_head bitmap instead of generic_head to track duplicates. gcc/cp/ * semantics.c (finish_omp_clauses): For C_ORT_OMP OMP_CLAUSE_USE_DEVICE_* clauses use oacc_reduction_head bitmap instead of generic_head to track duplicates. libgomp/ * target.c (gomp_map_vars_internal): For GOMP_MAP_USE_DEVICE_PTR perform the lookup in the first loop only if !not_found_cnt, otherwise perform lookups for it in the second loop guarded with if (not_found_cnt || has_firstprivate). * testsuite/libgomp.c/target-37.c: New test. * testsuite/libgomp.c++/target-22.C: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@274206 138bc75d-0d04-0410-961f-82ee72b054a4
2019-08-07 * tree-core.h (enum omp_clause_code): Adjust OMP_CLAUSE_USE_DEVICE_PTRjakub
OpenMP description. Add OMP_CLAUSE_USE_DEVICE_ADDR clause. * tree.c (omp_clause_num_ops, omp_clause_code_name): Add entries for OMP_CLAUSE_USE_DEVICE_ADDR clause. (walk_tree_1): Handle OMP_CLAUSE_USE_DEVICE_ADDR. * tree-pretty-print.c (dump_omp_clause): Likewise. * tree-nested.c (convert_nonlocal_omp_clauses, convert_local_omp_clauses): Likewise. * gimplify.c (gimplify_scan_omp_clauses, gimplify_adjust_omp_clauses): Likewise. * omp-low.c (scan_sharing_clauses, lower_omp_target): Likewise. Treat OMP_CLAUSE_USE_DEVICE_ADDR like OMP_CLAUSE_USE_DEVICE_PTR clause with array or reference to array types, no matter what type except for reference it has. gcc/c-family/ * c-pragma.h (enum pragma_omp_clause): Add PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR. Set PRAGMA_OACC_CLAUSE_USE_DEVICE equal to PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR instead of being a separate enumeration value. gcc/c/ * c-parser.c (c_parser_omp_clause_name): Parse use_device_addr clause. (c_parser_omp_clause_use_device_addr): New function. (c_parser_omp_all_clauses): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR. (OMP_TARGET_DATA_CLAUSE_MASK): Add PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR. (c_parser_omp_target_data): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR like PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR, adjust diagnostics about no map or use_device_* clauses. * c-typeck.c (c_finish_omp_clauses): For OMP_CLAUSE_USE_DEVICE_PTR in OpenMP, require pointer type rather than pointer or array type. Handle OMP_CLAUSE_USE_DEVICE_ADDR. gcc/cp/ * parser.c (cp_parser_omp_clause_name): Parse use_device_addr clause. (cp_parser_omp_all_clauses): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR. (OMP_TARGET_DATA_CLAUSE_MASK): Add PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR. (cp_parser_omp_target_data): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR like PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR, adjust diagnostics about no map or use_device_* clauses. * semantics.c (finish_omp_clauses): For OMP_CLAUSE_USE_DEVICE_PTR in OpenMP, require pointer or reference to pointer type rather than pointer or array or reference to pointer or array type. Handle OMP_CLAUSE_USE_DEVICE_ADDR. * pt.c (tsubst_omp_clauses): Handle OMP_CLAUSE_USE_DEVICE_ADDR. gcc/testsuite/ * c-c++-common/gomp/target-data-1.c (foo): Use use_device_addr clause instead of use_device_ptr clause where required by OpenMP 5.0, add further tests for both use_device_ptr and use_device_addr clauses. libgomp/ * testsuite/libgomp.c/target-18.c (struct S): New type. (foo): Use use_device_addr clause instead of use_device_ptr clause where required by OpenMP 5.0, add further tests for both use_device_ptr and use_device_addr clauses. * testsuite/libgomp.c++/target-9.C (struct S): New type. (foo): Use use_device_addr clause instead of use_device_ptr clause where required by OpenMP 5.0, add further tests for both use_device_ptr and use_device_addr clauses. Add t and u arguments. (main): Adjust caller. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@274159 138bc75d-0d04-0410-961f-82ee72b054a4
2019-08-06 * tree.h (OMP_CLAUSE_LASTPRIVATE_TASKLOOP_IV): Rename to ...jakub
(OMP_CLAUSE_LASTPRIVATE_LOOP_IV): ... this. Adjust comment. * gimplify.c (gimple_add_tmp_var): In SIMD contexts, turn addressable new vars into GOVD_PRIVATE rather than GOVD_LOCAL. (gimplify_omp_for): Don't do C++ random access iterator clause adjustments on combined constructs from OMP_LOOP. For OMP_LOOP, don't predetermine the artificial iterator in case of C++ random access iterators as lastprivate, but private. For OMP_LOOP, force bind expr around simd body and force for_pre_body before the construct. Use OMP_CLAUSE_LASTPRIVATE_LOOP_IV instead of OMP_CLAUSE_LASTPRIVATE_TASKLOOP_IV. (gimplify_omp_loop): Add firstprivate clauses on OMP_PARALLEL for diff var of C++ random access iterators. Handle OMP_CLAUSE_FIRSTPRIVATE. For OMP_CLAUSE_LASTPRIVATE_LOOP_IV, if not outermost also add OMP_CLAUSE_FIRSTPRIVATE, and in both cases clear OMP_CLAUSE_LASTPRIVATE_LOOP_IV on the lastprivate clause on the OMP_FOR and OMP_DISTRIBUTE constructs if any. * omp-low.c (lower_rec_input_clauses): For OMP_CLAUSE_LASTPRIVATE_LOOP_IV on simd copy construct the private variables instead of default constructing them. (lower_lastprivate_clauses): Use OMP_CLAUSE_LASTPRIVATE_LOOP_IV instead of OMP_CLAUSE_LASTPRIVATE_TASKLOOP_IV and move the is_taskloop_ctx check from the assert to the guarding condition. gcc/cp/ * parser.c (cp_parser_omp_for_loop): For OMP_LOOP, ignore parallel clauses and predetermine iterator as lastprivate. * semantics.c (handle_omp_for_class_iterator): Use OMP_CLAUSE_LASTPRIVATE_LOOP_IV instead of OMP_CLAUSE_LASTPRIVATE_TASKLOOP_IV, set it for lastprivate also on OMP_LOOP construct. If a clause is missing for class iterator on OMP_LOOP, add firstprivate clause, and if there is private clause, turn it into firstprivate too. (finish_omp_for): Formatting fix. For OMP_LOOP, adjust OMP_CLAUSE_LASTPRIVATE_LOOP_IV clause CP_CLAUSE_INFO, so that it uses copy ctor instead of default ctor. * cp-gimplify.c (cp_gimplify_expr): Handle OMP_LOOP like OMP_DISTRIBUTE etc. (cp_fold_r): Likewise. (cp_genericize_r): Likewise. (cxx_omp_finish_clause): Also finish lastprivate clause with OMP_CLAUSE_LASTPRIVATE_LOOP_IV flag. * pt.c (tsubst_omp_clauses): Handle OMP_CLAUSE_BIND. (tsubst_omp_for_iterator): For OMP_LOOP, ignore parallel clauses and predetermine iterator as lastprivate. * constexpr.c (potential_constant_expression_1): Handle OMP_LOOP like OMP_DISTRIBUTE etc. libgomp/ * testsuite/libgomp.c++/loop-13.C: New test. * testsuite/libgomp.c++/loop-14.C: New test. * testsuite/libgomp.c++/loop-15.C: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@274138 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-31 PR middle-end/91301jakub
* gimplify.c (gimplify_omp_for): If for class iterator on distribute parallel for there is no data sharing clause on inner_for_stmt, look for private clause on combined parallel too and if found, move it to inner_for_stmt. * testsuite/libgomp.c++/for-27.C: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@273922 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-232019-07-23 Steven G. Kargl <kargl@gcc.gnu.org>kargl
* arith.c (gfc_convert_integer, gfc_convert_real, gfc_convert_complex): Move to ... * primary.c (convert_integer, convert_real, convert_complex): ... here. Rename and make static functions. (match_integer_constant): Use convert_integer (match_real_constant): Use convert_real. (match_complex_constant: Use convert_complex. * arith.h (gfc_convert_integer, gfc_convert_real, gfc_convert_complex): Remove prototypes. * array.c (match_array_cons_element): A BOZ cannot be a data statement value. Jump to a common exit point. * check.c (gfc_invalid_boz): New function. Emit error or warning for a BOZ in an invalid context. (boz_args_check): Move to top of file to prevent need of forward declaration. (is_boz_constant): New function. Check that BOZ expr is constant. (gfc_b z2real): New function. In-place conversion of BOZ literal constant to REAL in accordance to F2018. (gfc_boz2int): New function. In-place conversion of BOZ literal onstant to INTEGER in accordance to F2018. (gfc_check_achar, gfc_check_char, gfc_check_float): Use gfc_invalid_boz. Convert BOZ as needed. (gfc_check_bge_bgt_ble_blt): Enforce F2018 requirements on BGE, BGT, BLE, and BLT intrinsic functions. (gfc_check_cmplx): Re-organize to check kind, if present, first. Convert BOZ real and/or imaginary parts as needed in accordance to F2018. (gfc_check_complex): Use gfc_invalid_boz. Convert BOZ as needed. (gfc_check_dcmplx, gfc_check_dble ): Convert BOZ as needed. (gfc_check_dshift): Make dshift[lr] conform to F2018 standard. gfc_check_float (gfc_expr *a) (gfc_check_iand_ieor_ior): Make IAND, IEOR, and IOR conform to F2018 standard. (gfc_check_int): Conform to F2018 standard. (gfc_check_intconv): Deprecate SHORT and LONG aliases for INT2 and INT. Simply return for a BOZ argument. See gfc_simplify_intconv. (gfc_check_merge_bits): Make MERGE_BITS conform to Fortran 2018 standard. (gfc_check_real): Remove incorrect comment. Check kind, if present, first. Simply return for a BOZ argument. See gfc_simplify_real. (gfc_check_and): Re-do error handling for BOZ arguments. Remove special casing ts.type != BT_INTEGER or BT_LOGICAL. * decl.c (match_old_style_init): Check for BOZ in old-style initialization. Issue error or warning depending on -fallow-invalid-boz option. Issue error if variable is not an INTEGER or REAL and the value is BOZ. * expr.c (gfc_copy_expr): Copy a BT_BOZ gfc_expr. (gfc_check_assign): Re-do error handling for a BOZ in an assignment statement. Do in-place conversion of RHS based on LHS type of INTEGER or REAL. * gfortran.h (gfc_expr): Add a boz component. Remove is_boz component. (gfc_boz2int, gfc_boz2real, gfc_invalid_boz): New prototypes. * interface.c (gfc_extend_assign): Guard against replacing an intrinsic involving a BOZ literal constant on RHS. * invoke.texi: Doument -fallow-invalid-boz. * lang.opt: New option. -fallow-invalid-boz. * libgfortran.h (bt): Elevate BOZ to a basic type. * misc.c (gfc_basic_typename, gfc_typename): Translate BT_BOZ to BOZ. * primary.c (convert_integer, convert_real, convert_complex): to here. Rename and make static functions. * primary.c(match_boz_constant): Rewrite parsing of a BOZ. Re-do error handling. Deprecate 'X' for hexidecimal and postfix notation. Use -fallow-invalid-boz and gfc_invalid_boz to accept deprecated code. * resolve.c (resolve_ordinary_assign): Rework a RHS that is a BOZ literal constant. Use gfc_invalid_boz to allow previous nonstandard behavior. Remove range checking of BOZ conversion. * simplify.c (convert_boz): Remove function. (simplify_cmplx): Remove conversion of BOZ constants, because conversion is done in gfc_check_cmplx. (gfc_simplify_float): Remove conversion of BOZ constant, because conversion is done in gfc_check_float. (simplify_intconv): Use gfc_boz2int to convert BOZ to INTEGER. Remove range checking for BOZ conversion. (gfc_simplify_real): Use k, if present, to determine kind. Convert BOZ to REAL. Remove range checking for BOZ conversion. target-memory.c (gfc_convert_boz): Rewrite to deal with convert of a BOZ to a REAL value. 2019-07-23 Steven G. Kargl <kargl@gcc.gnu.org> * gfortran.dg/achar_5.f90: Fix for new BOZ handling. * arithmetic_overflow_1.f90: Ditto. * gfortran.dg/boz_11.f90: Ditto. * gfortran.dg/boz_12.f90: Ditto. * gfortran.dg/boz_4.f90: Ditto. * gfortran.dg/boz_5.f90: Ditto. * gfortran.dg/boz_6.f90: Ditto. * gfortran.dg/boz_7.f90: Ditto. * gfortran.dg/boz_8.f90: Ditto. * gfortran.dg/dec_structure_6.f90: Ditto. * gfortran.dg/dec_union_1.f90: Ditto. * gfortran.dg/dec_union_2.f90: Ditto. * gfortran.dg/dec_union_5.f90: Ditto. * gfortran.dg/dshift_3.f90: Ditto. * gfortran.dg/gnu_logical_2.f90: Ditto. * gfortran.dg/int_conv_1.f90: Ditto. * gfortran.dg/ishft_1.f90: Ditto. * gfortran.dg/nan_4.f90: Ditto. * gfortran.dg/no_range_check_3.f90: Ditto. * gfortran.dg/pr16433.f: Ditto. * gfortran.dg/pr44491.f90: Ditto. * gfortran.dg/pr58027.f90: Ditto. * gfortran.dg/pr81509_2.f90: Ditto. * gfortran.dg/unf_io_convert_1.f90: Ditto. * gfortran.dg/unf_io_convert_2.f90: Ditto. * gfortran.fortran-torture/execute/intrinsic_fraction_exponent.f90: Ditto. * gfortran.fortran-torture/execute/intrinsic_mvbits.f90: Ditto. * gfortran.fortran-torture/execute/intrinsic_nearest.f90: Ditto. * gfortran.fortran-torture/execute/seq_io.f90: Ditto. * gfortran.dg/gnu_logical_1.F: Delete test. * gfortran.dg/merge_bits_3.f90: New test. * gfortran.dg/merge_bits_3.f90: Ditto. * gfortran.dg/boz_int.f90: Ditto. * gfortran.dg/boz_bge.f90: Ditto. * gfortran.dg/boz_complex_1.f90: Ditto. * gfortran.dg/boz_complex_2.f90: Ditto. * gfortran.dg/boz_complex_3.f90: Ditto. * gfortran.dg/boz_dble.f90: Ditto. * gfortran.dg/boz_dshift_1.f90: Ditto. * gfortran.dg/boz_dshift_2.f90: Ditto. * gfortran.dg/boz_float_1.f90: Ditto. * gfortran.dg/boz_float_2.f90: Ditto. * gfortran.dg/boz_float_3.f90: Ditto. * gfortran.dg/boz_iand_1.f90: Ditto. * gfortran.dg/boz_iand_2.f90: Ditto. 2019-07-23 Steven G. Kargl <kargl@gcc.gnu.org> * testsuite/libgomp.fortran/reduction4.f90: Update BOZ usage * testsuite/libgomp.fortran/reduction5.f90: Ditto. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@273747 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-20 * tree.def (OMP_LOOP): New tree code.jakub
* tree-core.h (enum omp_clause_code): Add OMP_CLAUSE_BIND. (enum omp_clause_bind_kind): New enum. (struct tree_omp_clause): Add subcode.bind_kind. * tree.h (OMP_LOOP_CHECK): Rename to ... (OMP_LOOPING_CHECK): ... this. (OMP_FOR_BODY, OMP_FOR_CLAUSES, OMP_FOR_INIT, OMP_FOR_COND, OMP_FOR_INCR, OMP_FOR_PRE_BODY, OMP_FOR_ORIG_DECLS): Use OMP_LOOPING_CHECK instead of OMP_LOOP_CHECK. (OMP_CLAUSE_BIND_KIND): Define. * tree.c (omp_clause_num_ops, omp_clause_code_name): Add bind clause entries. (walk_tree_1): Handle OMP_CLAUSE_BIND. * tree-pretty-print.c (dump_omp_clause): Likewise. (dump_generic_node): Handle OMP_LOOP. * gimplify.c (enum omp_region_type): Add ORT_IMPLICIT_TARGET. (in_omp_construct): New variable. (is_gimple_stmt): Handle OMP_LOOP. (gimplify_scan_omp_clauses): For lastprivate don't set check_non_private if code == OMP_LOOP. For reduction clause on OMP_LOOP combined with parallel or teams propagate as shared on the combined construct. Handle OMP_CLAUSE_BIND. (gimplify_adjust_omp_clauses): Handle OMP_CLAUSE_BIND. (gimplify_omp_for): Pass OMP_LOOP instead of OMP_{FOR,DISTRIBUTE} for constructs from a loop construct to gimplify_scan_omp_clauses. Don't predetermine iterator linear on OMP_SIMD from loop construct. (replace_reduction_placeholders, gimplify_omp_loop): New functions. (gimplify_omp_workshare): Use ORT_IMPLICIT_TARGET instead of trying to match the implicit ORT_TARGET construct around whole body. Temporarily clear in_omp_construct when processing body. (gimplify_expr): Handle OMP_LOOP. For OMP_MASTER, OMP_TASKGROUP etc. temporarily set in_omp_construct when processing body. (gimplify_body): Create ORT_IMPLICIT_TARGET instead of ORT_TARGET. * omp-low.c (struct omp_context): Add loop_p. (build_outer_var_ref): Treat ctx->loop_p similarly to simd construct in that the original var might be private. (scan_sharing_clauses): Handle OMP_CLAUSE_BIND. (check_omp_nesting_restrictions): Adjust nesting restrictions for addition of loop construct. (scan_omp_1_stmt): Allow setjmp inside of loop construct. gcc/c-family/ * c-pragma.h (enum pragma_kind): Add PRAGMA_OMP_LOOP. (enum pragma_omp_clause): Add PRAGMA_OMP_CLAUSE_BIND. * c-pragma.c (omp_pragmas_simd): Add PRAGMA_OMP_LOOP entry. * c-common.h (enum c_omp_clause_split): Add C_OMP_CLAUSE_SPLIT_LOOP. * c-omp.c (c_omp_split_clauses): Add support for 4 new combined constructs with the loop construct. gcc/c/ * c-parser.c (c_parser_omp_clause_name): Handle bind clause. (c_parser_omp_clause_bind): New function. (c_parser_omp_all_clauses): Handle PRAGMA_OMP_CLAUSE_BIND. (OMP_LOOP_CLAUSE_MASK): Define. (c_parser_omp_loop): New function. (c_parser_omp_parallel, c_parser_omp_teams): Handle parsing of loop combined with parallel or teams. (c_parser_omp_construct): Handle PRAGMA_OMP_LOOP. * c-typeck.c (c_finish_omp_clauses): Handle OMP_CLAUSE_BIND. gcc/cp/ * cp-tree.h (OMP_FOR_GIMPLIFYING_P): Use OMP_LOOPING_CHECK instead of OMP_LOOP_CHECK. * parser.c (cp_parser_omp_clause_name): Handle bind clause. (cp_parser_omp_clause_bind): New function. (cp_parser_omp_all_clauses): Handle PRAGMA_OMP_CLAUSE_BIND. (OMP_LOOP_CLAUSE_MASK): Define. (cp_parser_omp_loop): New function. (cp_parser_omp_parallel, cp_parser_omp_teams): Handle parsing of loop combined with parallel or teams. (cp_parser_omp_construct): Handle PRAGMA_OMP_LOOP. (cp_parser_pragma): Likewise. * pt.c (tsubst_expr): Handle OMP_LOOP. * semantics.c (finish_omp_clauses): Handle OMP_CLAUSE_BIND. gcc/testsuite/ * c-c++-common/gomp/cancel-1.c: Adjust expected diagnostic wording. * c-c++-common/gomp/clauses-1.c (foo, baz, bar): Add order(concurrent) clause where allowed. Add combined constructs with loop with all possible clauses. (qux): New function. * c-c++-common/gomp/loop-1.c: New test. * c-c++-common/gomp/loop-2.c: New test. * c-c++-common/gomp/loop-3.c: New test. * c-c++-common/gomp/loop-4.c: New test. * c-c++-common/gomp/loop-5.c: New test. * c-c++-common/gomp/order-3.c: Adjust expected diagnostic wording. * c-c++-common/gomp/simd-setjmp-1.c: New test. * c-c++-common/gomp/teams-2.c: Adjust expected diagnostic wording. libgomp/ * testsuite/libgomp.c-c++-common/loop-1.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@273621 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-08 * tree-vect-stmts.c (scan_operand_equal_p): Look through MEM_REFjakub
with SSA_NAME address of POINTER_PLUS_EXPR. Handle MULT_EXPR and casts in offset when different, both through gimple stmts and through trees. Rewritten using loops to minimize code duplication for each operand. * g++.dg/vect/simd-6.cc: Replace xfail with target x86. * g++.dg/vect/simd-9.cc: Likewise. * testsuite/libgomp.c++/scan-13.C: Replace xfail with target x86. * testsuite/libgomp.c++/scan-16.C: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@273249 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-06 * omp-low.c (lower_rec_input_clauses): For lastprivate clauses injakub
ctx->for_simd_scan_phase simd copy the outer var to the privatized variable(s). For conditional lastprivate look through outer GIMPLE_OMP_SCAN context. (lower_omp_1): For conditional lastprivate look through outer GIMPLE_OMP_SCAN context. * testsuite/libgomp.c/scan-19.c: New test. * testsuite/libgomp.c/scan-20.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@273169 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-06 * omp-low.c (struct omp_context): Add for_simd_scan_phase member.jakub
(maybe_lookup_ctx): Add forward declaration. (omp_find_scan): Likewise. Walk into body of simd if composited with worksharing loop. (scan_omp_simd_scan): New function. (scan_omp_1_stmt): Call it. (lower_rec_simd_input_clauses): Don't create rvar nor rvar2 if ctx->for_simd_scan_phase. (lower_rec_input_clauses): Do much less work for inscan reductions in ctx->for_simd_scan_phase is_simd regions. (lower_omp_scan): Set is_simd also on simd constructs composited with worksharing loop, unless ctx->for_simd_scan_phase. Never emit a sorry message. Don't change GIMPLE_OMP_SCAN stmts into nops and emit their body after in simd constructs composited with worksharing loop. (lower_omp_for_scan): Handle worksharing loop composited with simd. * c-c++-common/gomp/scan-4.c: Don't expect sorry message. * testsuite/libgomp.c/scan-11.c: New test. * testsuite/libgomp.c/scan-12.c: New test. * testsuite/libgomp.c/scan-13.c: New test. * testsuite/libgomp.c/scan-14.c: New test. * testsuite/libgomp.c/scan-15.c: New test. * testsuite/libgomp.c/scan-16.c: New test. * testsuite/libgomp.c/scan-17.c: New test. * testsuite/libgomp.c/scan-18.c: New test. * testsuite/libgomp.c++/scan-9.C: New test. * testsuite/libgomp.c++/scan-10.C: New test. * testsuite/libgomp.c++/scan-11.C: New test. * testsuite/libgomp.c++/scan-12.C: New test. * testsuite/libgomp.c++/scan-13.C: New test. * testsuite/libgomp.c++/scan-14.C: New test. * testsuite/libgomp.c++/scan-15.C: New test. * testsuite/libgomp.c++/scan-16.C: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@273157 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-04 * omp-expand.c (expand_omp_for_static_nochunk): Don't emitjakub
GOMP_loop_start at the start of second worksharing loop in a scan. For nowait, don't emit GOMP_loop_end_nowait at the end of first worksharing loop in a scan even if there are conditional lastprivates, and do emit GOMP_loop_end_nowait at the end of second worksharing loop. * testsuite/libgomp.c/scan-9.c: New test. * testsuite/libgomp.c/scan-10.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@273095 138bc75d-0d04-0410-961f-82ee72b054a4
2019-07-03 * tree-core.h (enum omp_clause_code): Add OMP_CLAUSE__SCANTEMP_jakub
clause. * tree.h (OMP_CLAUSE_DECL): Use OMP_CLAUSE__SCANTEMP_ instead of OMP_CLAUSE__CONDTEMP_ as range's upper bound. (OMP_CLAUSE__SCANTEMP__ALLOC, OMP_CLAUSE__SCANTEMP__CONTROL): Define. * tree.c (omp_clause_num_ops, omp_clause_code_name): Add OMP_CLAUSE__SCANTEMP_ entry. (walk_tree_1): Handle OMP_CLAUSE__SCANTEMP_. * tree-pretty-print.c (dump_omp_clause): Likewise. * tree-nested.c (convert_nonlocal_omp_clauses, convert_local_omp_clauses): Likewise. * omp-general.h (struct omp_for_data): Add have_scantemp and have_nonctrl_scantemp members. * omp-general.c (omp_extract_for_data): Initialize them. * omp-low.c (struct omp_context): Add scan_exclusive member. (scan_omp_1_stmt): Don't unnecessarily mask gimple_omp_for_kind result again with GF_OMP_FOR_KIND_MASK. Initialize also ctx->scan_exclusive. (lower_rec_simd_input_clauses): Use ctx->scan_exclusive instead of !ctx->scan_inclusive. (lower_rec_input_clauses): Simplify gimplification of dtors using gimplify_and_add. For non-is_simd test OMP_CLAUSE_REDUCTION_INSCAN rather than rvarp. Handle OMP_CLAUSE_REDUCTION_INSCAN in worksharing loops. Don't add barrier for reduction_omp_orig_ref if ctx->scan_??xclusive. (lower_reduction_clauses): Don't do anything for ctx->scan_??xclusive. (lower_omp_scan): Use ctx->scan_exclusive instead of !ctx->scan_inclusive. Handle worksharing loops with inscan reductions. Use new_vard != new_var instead of repeated omp_is_reference calls. (omp_find_scan, lower_omp_for_scan): New functions. (lower_omp_for): Call lower_omp_for_scan for worksharing loops with inscan reductions. * omp-expand.c (expand_omp_scantemp_alloc): New function. (expand_omp_for_static_nochunk): Handle fd->have_nonctrl_scantemp and fd->have_scantemp. * c-c++-common/gomp/scan-3.c (f1): Don't expect a sorry message. * c-c++-common/gomp/scan-5.c (foo): Likewise. * testsuite/libgomp.c++/scan-1.C: New test. * testsuite/libgomp.c++/scan-2.C: New test. * testsuite/libgomp.c++/scan-3.C: New test. * testsuite/libgomp.c++/scan-4.C: New test. * testsuite/libgomp.c++/scan-5.C: New test. * testsuite/libgomp.c++/scan-6.C: New test. * testsuite/libgomp.c++/scan-7.C: New test. * testsuite/libgomp.c++/scan-8.C: New test. * testsuite/libgomp.c/scan-1.c: New test. * testsuite/libgomp.c/scan-2.c: New test. * testsuite/libgomp.c/scan-3.c: New test. * testsuite/libgomp.c/scan-4.c: New test. * testsuite/libgomp.c/scan-5.c: New test. * testsuite/libgomp.c/scan-6.c: New test. * testsuite/libgomp.c/scan-7.c: New test. * testsuite/libgomp.c/scan-8.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272958 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-18Test cases to verify OpenACC 'firstprivate' mappingstschwinge
gcc/testsuite/ * c-c++-common/goacc/firstprivate-mappings-1.c: New file. * g++.dg/goacc/firstprivate-mappings-1.C: Likewise. libgomp/ * testsuite/libgomp.oacc-c++/firstprivate-mappings-1.C: New file. * testsuite/libgomp.oacc-c-c++-common/firstprivate-mappings-1.c: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272451 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-18Add missing results check in 'libgomp.fortran/allocatable3.f90'tschwinge
libgomp/ * testsuite/libgomp.fortran/allocatable3.f90: Add missing results check. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272449 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-18Add 'libgomp.oacc-fortran/allocatable-array-1.f90'tschwinge
libgomp/ * testsuite/libgomp.oacc-fortran/allocatable-array-1.f90: New file. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272448 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-18[PR90743] Fortran 'allocatable' with OpenACC data/OpenMP 'target' 'map' clausestschwinge
Test what OpenMP 5.0 has to say on this topic. And, do the same for OpenACC. libgomp/ PR fortran/90743 * oacc-parallel.c (GOACC_parallel_keyed): Handle NULL mapping case. * testsuite/libgomp.fortran/target-allocatable-1-1.f90: New file. * testsuite/libgomp.fortran/target-allocatable-1-2.f90: Likewise. * testsuite/libgomp.oacc-fortran/allocatable-1-1.f90: Likewise. * testsuite/libgomp.oacc-fortran/allocatable-1-2.f90: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272447 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-18[PR90861] Document status quo for OpenACC 'declare' not cleaning up for VLAstschwinge
gcc/testsuite/ PR testsuite/90861 * c-c++-common/goacc/declare-pr90861.c: New file. libgomp/ PR testsuite/90861 * testsuite/libgomp.oacc-c-c++-common/declare-vla.c: Update. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272446 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-18[PR90862] OpenACC 'declare' ICE when nested inside another constructtschwinge
gcc/ PR middle-end/90862 * omp-low.c (check_omp_nesting_restrictions): Handle GF_OMP_TARGET_KIND_OACC_DECLARE. gcc/testsuite/ PR middle-end/90862 * c-c++-common/goacc/declare-1.c: Update. * c-c++-common/goacc/declare-2.c: Likewise. libgomp/ PR middle-end/90862 * testsuite/libgomp.oacc-c-c++-common/declare-1.c: Update. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272444 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-16[openacc, parloops] Fix SIGSEGV in oacc_entry_exit_ok_1vries
When compiling the test-case with r268755, we run into a SIGSEGV in oacc_entry_exit_ok_1 when trying to dereference a NULL red: ... struct reduction_info *red; red = reduction_phi (reduction_list, use_stmt); tree val = PHI_RESULT (red->keep_res); ... Fix this by handling ref == NULL. Bootstrapped and reg-tested on x86_64. Build and reg-tested on x86_64 with nvptx accelerator. 2019-06-16 Tom de Vries <tdevries@suse.de> PR tree-optimization/89376 * tree-parloops.c (oacc_entry_exit_ok_1): Handle red == NULL. * testsuite/libgomp.oacc-c-c++-common/pr89376.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272338 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-15[nvptx, libgomp] Update pr85381-{2,4}.c test-casesvries
After the fix for "PR tree-optimization/89713 - Assume loop with an exit is finite" ( r272234 ) empty oacc loops are removed before expand. Update pr85381-{2,4}.c accordingly. 2019-06-15 Tom de Vries <tdevries@suse.de> PR tree-optimization/89713 * testsuite/libgomp.oacc-c-c++-common/pr85381-2.c: Expect no bar.sync. * testsuite/libgomp.oacc-c-c++-common/pr85381-4.c: Same. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272324 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-15 PR middle-end/90779jakub
* gimplify.c: Include omp-offload.h and context.h. (gimplify_bind_expr): Add "omp declare target" attributes to static block scope variables inside of target region or target functions. * c-c++-common/goacc/routine-5.c (func2): Don't expect error for static block scope variable in #pragma acc routine. * testsuite/libgomp.c/pr90779.c: New test. * testsuite/libgomp.fortran/pr90779.f90: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272322 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-15[openacc] Disable pass_thread_jumps for IFN_UNIQUEvries
If we compile the openacc testcase with -fopenacc -O2, we run into a SIGSEGV or assert. The root cause for this is that pass_thread_jumps breaks the invariant that OACC_FORK and OACC_JOIN mark the start and end of a single-entry-single-exit region. Fix this by bailing out when encountering an IFN_UNIQUE in thread_jumps::profitable_jump_thread_path. Bootstrapped and reg-tested on x86_64. Build and reg-tested libgomp on x86_64 with nvptx accelerator. 2019-06-15 Tom de Vries <tdevries@suse.de> PR tree-optimization/90009 * tree-ssa-threadbackward.c (thread_jumps::profitable_jump_thread_path): Return NULL if bb contains IFN_UNIQUE. * testsuite/libgomp.oacc-c-c++-common/pr90009.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272321 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-13PR tree-optimization/89713 - Assume loop with an exit is finitefxue
gcc/ChangeLog: * doc/invoke.texi (-ffinite-loops): Document new option. * common.opt (-ffinite-loops): New option. * tree-ssa-dce.c (mark_stmt_if_obviously_necessary): Mark IFN_GOACC_LOOP calls as necessary. * tree-ssa-loop-niter.c (finite_loop_p): Assume loop with an exit is finite. * omp-offload.c (oacc_xform_loop): Skip lowering if return value of IFN_GOACC_LOOP call is not used. * opts.c (default_options_table): Enable -ffinite-loops at -O2+. gcc/testsuite/ChangeLog: * g++.dg/tree-ssa/empty-loop.C: New test. * gcc.dg/tree-ssa/dce-2.c: New test. * gcc.dg/const-1.c: Add -fno-finite-loops option. * gcc.dg/graphite/graphite.exp: Likewise. * gcc.dg/loop-unswitch-1.c: Likewise. * gcc.dg/predict-9.c: Likewise. * gcc.dg/pure-2.c: Likewise. * gcc.dg/tree-ssa/20040211-1.c: Likewise. * gcc.dg/tree-ssa/loop-10.c: Likewise. * gcc.dg/tree-ssa/split-path-6.c: Likewise. * gcc.dg/tree-ssa/ssa-thread-12.c: Likewise. libgomp/ChangeLog: * testsuite/libgomp.oacc-c-c++-common/pr84955-1.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272234 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-11 PR target/90811jakub
* config/nvptx/nvptx.c (nvptx_output_softstack_switch): Use and.b%d instead of and.u%d. * testsuite/libgomp.c/pr90811.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@272161 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-05 * omp-low.c (lower_rec_input_clauses): For lastprivate conditionaljakub
references, lookup in in hash map MEM_REF operand instead of the MEM_REF itself. (lower_omp_1): When looking for lastprivate conditional assignments, handle MEM_REFs with REFERENCE_TYPE operands. * testsuite/libgomp.c++/lastprivate-conditional-1.C: New test. * testsuite/libgomp.c++/lastprivate-conditional-2.C: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271948 138bc75d-0d04-0410-961f-82ee72b054a4
2019-06-04 * gimplify.c (gimplify_scan_omp_clauses): Don't sorry_at on lastprivatejakub
conditional on combined for simd. * omp-low.c (struct omp_context): Add combined_into_simd_safelen0 member. (lower_rec_input_clauses): For gimple_omp_for_combined_into_p max_vf 1 constructs, don't remove lastprivate_conditional_map, but instead set ctx->combined_into_simd_safelen0 and adjust hash_map, so that it points to parent construct temporaries. (lower_lastprivate_clauses): Handle ctx->combined_into_simd_safelen0 like !ctx->lastprivate_conditional_map. (lower_omp_1) <case GIMPLE_ASSIGN>: If up->combined_into_simd_safelen0, use up->outer context instead of up. * omp-expand.c (expand_omp_for_generic): Perform cond_var bump even if gimple_omp_for_combined_p. (expand_omp_for_static_nochunk): Likewise. (expand_omp_for_static_chunk): Add forgotten cond_var bump that was probably moved over into expand_omp_for_generic rather than being copied there. gcc/cp/ * cp-tree.h (CP_OMP_CLAUSE_INFO): Allow for any clauses up to _condvar_ instead of only up to linear. gcc/testsuite/ * c-c++-common/gomp/lastprivate-conditional-2.c (foo): Don't expect a sorry_at on any of the clauses. libgomp/ * testsuite/libgomp.c-c++-common/lastprivate-conditional-7.c: New test. * testsuite/libgomp.c-c++-common/lastprivate-conditional-8.c: New test. * testsuite/libgomp.c-c++-common/lastprivate-conditional-9.c: New test. * testsuite/libgomp.c-c++-common/lastprivate-conditional-10.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271907 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-30Generalize getconf _NPROCESSORS_ONLNro
libgomp: * configure.ac: Call AX_COUNT_CPUS. Substitute CPU_COUNT. * testsuite/Makefile.am (check-am): Use CPU_COUNT as processor count fallback. * aclocal.m4: Regenerate. * configure: Regenerate. * Makefile.in, testsuite/Makefile.in: Regenerate. config: * ax_count_cpus.m4: New file. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271769 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-29 * gimplify.c (struct gimplify_omp_ctx): Add clauses member.jakub
(gimplify_scan_omp_clauses): Initialize ctx->clauses. (gimplify_adjust_omp_clauses_1): Transform lastprivate conditional explicit clause on combined parallel into implicit shared clause. (gimplify_adjust_omp_clauses): Move lastprivate conditional clause and firstprivate if the decl has one too from combined parallel to the worksharing construct. gcc/testsuite/ * c-c++-common/gomp/lastprivate-conditional-2.c (foo): Don't expect sorry on lastprivate conditional on parallel for. * c-c++-common/gomp/lastprivate-conditional-3.c (foo): Add tests for lastprivate conditional warnings on parallel for constructs. * c-c++-common/gomp/lastprivate-conditional-4.c: New test. libgomp/ * testsuite/libgomp.c-c++-common/lastprivate_conditional_4.c: Rename to ... * testsuite/libgomp.c-c++-common/lastprivate-conditional-4.c: ... this. * testsuite/libgomp.c-c++-common/lastprivate-conditional-5.c: New test. * testsuite/libgomp.c-c++-common/lastprivate-conditional-6.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271733 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-27 * gimplify.c (gimplify_scan_omp_clauses): Allow lastprivate conditionaljakub
on sections construct. * omp-low.c (lower_lastprivate_conditional_clauses): Handle sections construct. (lower_omp_sections): Handle lastprivate conditional. (lower_omp_1) <case GIMPLE_ASSIGN>: Handle sections construct with lastprivate_conditional_map. * omp-expand.c (expand_omp_sections): Handle lastprivate conditional. libgomp/ * testsuite/libgomp.c-c++-common/lastprivate_conditional_4.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271673 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-27 * omp-low.c (lower_omp_1) <case GIMPLE_ASSIGN>: Look through ordered,jakub
critical, taskgroup and section regions when looking for a region with non-NULL lastprivate_conditional_map. * testsuite/libgomp.c-c++-common/lastprivate-conditional-3.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271672 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-27 PR libgomp/90641jakub
* work.c (gomp_init_work_share): Instead of aligning final ordered value to multiples of long long alignment, align to that the first part (ordered team ids) and if inline_ordered_team_ids is not on a long long alignment boundary within the structure, use __alignof__ (long long) - 1 pad size always. * loop.c (GOMP_loop_start): Fix *mem computation if inline_ordered_team_ids is not aligned on long long alignment boundary within the structure. * loop-ull.c (GOMP_loop_ull_start): Likewise. * sections.c (GOMP_sections2_start): Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271671 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-24 * tree-core.h (enum omp_clause_code): Add OMP_CLAUSE__CONDTEMP_.jakub
* tree.h (OMP_CLAUSE_DECL): Use OMP_CLAUSE__CONDTEMP_ instead of OMP_CLAUSE__REDUCTEMP_. * tree.c (omp_clause_num_ops, omp_clause_code_name): Add OMP_CLAUSE__CONDTEMP_. (walk_tree_1): Handle OMP_CLAUSE__CONDTEMP_. * tree-pretty-print.c (dump_omp_clause): Likewise. * tree-nested.c (convert_nonlocal_omp_clauses, convert_local_omp_clauses): Likewise. * gimplify.c (enum gimplify_omp_var_data): Use hexadecimal constants instead of decimal. Add GOVD_LASTPRIVATE_CONDITIONAL. (gimplify_scan_omp_clauses): Don't reject lastprivate conditional on OMP_FOR. (gimplify_omp_for): Warn and disable conditional modifier from lastprivate on loop iterators. * omp-general.h (struct omp_for_data): Add lastprivate_conditional member. * omp-general.c (omp_extract_for_data): Initialize it. * omp-low.c (struct omp_context): Add lastprivate_conditional_map member. (delete_omp_context): Delete it. (lower_lastprivate_conditional_clauses): New function. (lower_lastprivate_clauses): Add BODY_P and CSTMT_LIST arguments, handle lastprivate conditional clauses. (lower_reduction_clauses): Add CLIST argument, emit it into the critical section if any. (lower_omp_sections): Adjust lower_lastprivate_clauses and lower_reduction_clauses callers. (lower_omp_for_lastprivate): Add CLIST argument, pass it through to lower_lastprivate_clauses. (lower_omp_for): Call lower_lastprivate_conditional_clauses, adjust lower_omp_for_lastprivate and lower_reduction_clauses callers, emit clist into a critical section if not emitted there already by lower_reduction_clauses. (lower_omp_taskreg, lower_omp_teams): Adjust lower_reduction_clauses callers. (lower_omp_1): Handle GIMPLE_ASSIGNs storing into lastprivate conditional variables. * omp-expand.c (determine_parallel_type): Punt if OMP_CLAUSE__CONDTEMP_ clause is present. (expand_omp_for_generic, expand_omp_for_static_nochunk, expand_omp_for_static_chunk): Handle lastprivate conditional. (expand_omp_for): Handle fd.lastprivate_conditional like fd.have_reductemp. gcc/testsuite/ * c-c++-common/gomp/lastprivate-conditional-2.c (foo): Don't expect sorry for omp for. * c-c++-common/gomp/lastprivate-conditional-3.c: New test. libgomp/ * testsuite/libgomp.c-c++-common/lastprivate-conditional-1.c: New test. * testsuite/libgomp.c-c++-common/lastprivate-conditional-2.c: New test. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271610 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-24 PR libgomp/90585jakub
* plugin/plugin-hsa.c: Include gstdint.h. Include inttypes.h only if HAVE_INTTYPES_H is defined. (print_uint64_t): New typedef. (PRIu64): Define if HAVE_INTTYPES_H is not defined. (print_kernel_dispatch, run_kernel): Use PRIu64 macro instead of "lu", cast uint64_t HSA_DEBUG and fprintf arguments to print_uint64_t. (release_kernel_dispatch): Likewise. Cast shadow->debug to uintptr_t before casting to void *. * plugin/plugin-nvptx.c: Include gstdint.h instead of stdint.h. * oacc-mem.c: Don't include config.h nor stdint.h. * target.c: Don't include config.h. * oacc-cuda.c: Likewise. * oacc-host.c: Don't include stdint.h. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271597 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-20 PR libgomp/90527jakub
* alloc.c (_GNU_SOURCE): Define. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271438 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-17OpenACC Profiling Interface (incomplete)tschwinge
libgomp/ * acc_prof.h: New file. * oacc-profiling.c: Likewise. * Makefile.am (nodist_libsubinclude_HEADERS, libgomp_la_SOURCES): Add these, respectively. * Makefile.in: Regenerate. * env.c (initialize_env): Call goacc_profiling_initialize. * oacc-plugin.c (GOMP_PLUGIN_goacc_thread) (GOMP_PLUGIN_goacc_profiling_dispatch): New functions. * oacc-plugin.h (GOMP_PLUGIN_goacc_thread) (GOMP_PLUGIN_goacc_profiling_dispatch): Declare. * libgomp.map (OACC_2.5.1): Add acc_prof_lookup, acc_prof_register, acc_prof_unregister, and acc_register_library. (GOMP_PLUGIN_1.3): Add GOMP_PLUGIN_goacc_profiling_dispatch, and GOMP_PLUGIN_goacc_thread. * oacc-int.h (struct goacc_thread): Add prof_info, api_info, prof_callbacks_enabled members. (goacc_prof_enabled, goacc_profiling_initialize) (_goacc_profiling_dispatch_p, _goacc_profiling_setup_p) (goacc_profiling_dispatch): Declare. (GOACC_PROF_ENABLED, GOACC_PROFILING_DISPATCH_P) (GOACC_PROFILING_SETUP_P): Define. * oacc-async.c (acc_async_test, acc_async_test_all, acc_wait) (acc_wait_async, acc_wait_all, acc_wait_all_async): Update for OpenACC Profiling Interface. * oacc-cuda.c (acc_get_current_cuda_device) (acc_get_current_cuda_context, acc_get_cuda_stream) (acc_set_cuda_stream): Likewise. * oacc-init.c (acc_init_1, goacc_attach_host_thread_to_device) (acc_init, acc_set_device_type, acc_get_device_type) (acc_get_device_num, goacc_lazy_initialize): Likewise. * oacc-mem.c (acc_malloc, acc_free, memcpy_tofrom_device) (acc_deviceptr, acc_hostptr, acc_is_present, acc_map_data) (acc_unmap_data, present_create_copy, delete_copyout) (update_dev_host): Likewise. * oacc-parallel.c (GOACC_parallel_keyed, GOACC_data_start) (GOACC_data_end, GOACC_enter_exit_data, GOACC_update, GOACC_wait): Likewise. * plugin/plugin-nvptx.c (nvptx_exec, nvptx_alloc, nvptx_free) (GOMP_OFFLOAD_openacc_exec, GOMP_OFFLOAD_openacc_async_exec): Likewise. * libgomp.texi: Update. * testsuite/libgomp.oacc-c-c++-common/acc_prof-dispatch-1.c: New file. * testsuite/libgomp.oacc-c-c++-common/acc_prof-init-1.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_prof-kernels-1.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_prof-parallel-1.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_prof-valid_bytes-1.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_prof-version-1.c: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271346 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-132019-05-13 Chung-Lin Tang <cltang@codesourcery.com>cltang
Reviewed-by: Thomas Schwinge <thomas@codesourcery.com> libgomp/ * libgomp-plugin.h (struct goacc_asyncqueue): Declare. (struct goacc_asyncqueue_list): Likewise. (goacc_aq): Likewise. (goacc_aq_list): Likewise. (GOMP_OFFLOAD_openacc_register_async_cleanup): Remove. (GOMP_OFFLOAD_openacc_async_test): Remove. (GOMP_OFFLOAD_openacc_async_test_all): Remove. (GOMP_OFFLOAD_openacc_async_wait): Remove. (GOMP_OFFLOAD_openacc_async_wait_async): Remove. (GOMP_OFFLOAD_openacc_async_wait_all): Remove. (GOMP_OFFLOAD_openacc_async_wait_all_async): Remove. (GOMP_OFFLOAD_openacc_async_set_async): Remove. (GOMP_OFFLOAD_openacc_exec): Adjust declaration. (GOMP_OFFLOAD_openacc_cuda_get_stream): Likewise. (GOMP_OFFLOAD_openacc_cuda_set_stream): Likewise. (GOMP_OFFLOAD_openacc_async_exec): Declare. (GOMP_OFFLOAD_openacc_async_construct): Declare. (GOMP_OFFLOAD_openacc_async_destruct): Declare. (GOMP_OFFLOAD_openacc_async_test): Declare. (GOMP_OFFLOAD_openacc_async_synchronize): Declare. (GOMP_OFFLOAD_openacc_async_serialize): Declare. (GOMP_OFFLOAD_openacc_async_queue_callback): Declare. (GOMP_OFFLOAD_openacc_async_host2dev): Declare. (GOMP_OFFLOAD_openacc_async_dev2host): Declare. * libgomp.h (struct acc_dispatch_t): Define 'async' sub-struct. (gomp_acc_insert_pointer): Adjust declaration. (gomp_copy_host2dev): New declaration. (gomp_copy_dev2host): Likewise. (gomp_map_vars_async): Likewise. (gomp_unmap_tgt): Likewise. (gomp_unmap_vars_async): Likewise. (gomp_fini_device): Likewise. * oacc-async.c (get_goacc_thread): New function. (get_goacc_thread_device): New function. (lookup_goacc_asyncqueue): New function. (get_goacc_asyncqueue): New function. (acc_async_test): Adjust code to use new async design. (acc_async_test_all): Likewise. (acc_wait): Likewise. (acc_wait_async): Likewise. (acc_wait_all): Likewise. (acc_wait_all_async): Likewise. (goacc_async_free): New function. (goacc_init_asyncqueues): Likewise. (goacc_fini_asyncqueues): Likewise. * oacc-cuda.c (acc_get_cuda_stream): Adjust code to use new async design. (acc_set_cuda_stream): Likewise. * oacc-host.c (host_openacc_exec): Adjust parameters, remove 'async'. (host_openacc_register_async_cleanup): Remove. (host_openacc_async_exec): New function. (host_openacc_async_test): Adjust parameters. (host_openacc_async_test_all): Remove. (host_openacc_async_wait): Remove. (host_openacc_async_wait_async): Remove. (host_openacc_async_wait_all): Remove. (host_openacc_async_wait_all_async): Remove. (host_openacc_async_set_async): Remove. (host_openacc_async_synchronize): New function. (host_openacc_async_serialize): New function. (host_openacc_async_host2dev): New function. (host_openacc_async_dev2host): New function. (host_openacc_async_queue_callback): New function. (host_openacc_async_construct): New function. (host_openacc_async_destruct): New function. (struct gomp_device_descr host_dispatch): Remove initialization of old interface, add intialization of new async sub-struct. * oacc-init.c (acc_shutdown_1): Adjust to use gomp_fini_device. (goacc_attach_host_thread_to_device): Remove old async code usage. * oacc-int.h (goacc_init_asyncqueues): New declaration. (goacc_fini_asyncqueues): Likewise. (goacc_async_copyout_unmap_vars): Likewise. (goacc_async_free): Likewise. (get_goacc_asyncqueue): Likewise. (lookup_goacc_asyncqueue): Likewise. * oacc-mem.c (memcpy_tofrom_device): Adjust code to use new async design. (present_create_copy): Adjust code to use new async design. (delete_copyout): Likewise. (update_dev_host): Likewise. (gomp_acc_insert_pointer): Add async parameter, adjust code to use new async design. (gomp_acc_remove_pointer): Adjust code to use new async design. * oacc-parallel.c (GOACC_parallel_keyed): Adjust code to use new async design. (GOACC_enter_exit_data): Likewise. (goacc_wait): Likewise. (GOACC_update): Likewise. * oacc-plugin.c (GOMP_PLUGIN_async_unmap_vars): Change to assert fail when called, warn as obsolete in comment. * target.c (goacc_device_copy_async): New function. (gomp_copy_host2dev): Remove 'static', add goacc_asyncqueue parameter, add goacc_device_copy_async case. (gomp_copy_dev2host): Likewise. (gomp_map_vars_existing): Add goacc_asyncqueue parameter, adjust code. (gomp_map_pointer): Likewise. (gomp_map_fields_existing): Likewise. (gomp_map_vars_internal): New always_inline function, renamed from gomp_map_vars. (gomp_map_vars): Implement by calling gomp_map_vars_internal. (gomp_map_vars_async): Implement by calling gomp_map_vars_internal, passing goacc_asyncqueue argument. (gomp_unmap_tgt): Remove static, add attribute_hidden. (gomp_unref_tgt): New function. (gomp_unmap_vars_internal): New always_inline function, renamed from gomp_unmap_vars. (gomp_unmap_vars): Implement by calling gomp_unmap_vars_internal. (gomp_unmap_vars_async): Implement by calling gomp_unmap_vars_internal, passing goacc_asyncqueue argument. (gomp_fini_device): New function. (gomp_exit_data): Adjust gomp_copy_dev2host call. (gomp_load_plugin_for_device): Remove old interface, adjust to load new async interface. (gomp_target_fini): Adjust code to call gomp_fini_device. * plugin/plugin-nvptx.c (struct cuda_map): Remove. (struct ptx_stream): Remove. (struct nvptx_thread): Remove current_stream field. (cuda_map_create): Remove. (cuda_map_destroy): Remove. (map_init): Remove. (map_fini): Remove. (map_pop): Remove. (map_push): Remove. (struct goacc_asyncqueue): Define. (struct nvptx_callback): Define. (struct ptx_free_block): Define. (struct ptx_device): Remove null_stream, active_streams, async_streams, stream_lock, and next fields. (enum ptx_event_type): Remove. (struct ptx_event): Remove. (ptx_event_lock): Remove. (ptx_events): Remove. (init_streams_for_device): Remove. (fini_streams_for_device): Remove. (select_stream_for_async): Remove. (nvptx_init): Remove ptx_events and ptx_event_lock references. (nvptx_attach_host_thread_to_device): Remove CUDA_ERROR_NOT_PERMITTED case. (nvptx_open_device): Add free_blocks initialization, remove init_streams_for_device call. (nvptx_close_device): Remove fini_streams_for_device call, add free_blocks destruct code. (event_gc): Remove. (event_add): Remove. (nvptx_exec): Adjust parameters and code. (nvptx_free): Likewise. (nvptx_host2dev): Remove. (nvptx_dev2host): Remove. (nvptx_set_async): Remove. (nvptx_async_test): Remove. (nvptx_async_test_all): Remove. (nvptx_wait): Remove. (nvptx_wait_async): Remove. (nvptx_wait_all): Remove. (nvptx_wait_all_async): Remove. (nvptx_get_cuda_stream): Remove. (nvptx_set_cuda_stream): Remove. (GOMP_OFFLOAD_alloc): Adjust code. (GOMP_OFFLOAD_free): Likewise. (GOMP_OFFLOAD_openacc_register_async_cleanup): Remove. (GOMP_OFFLOAD_openacc_exec): Adjust parameters and code. (GOMP_OFFLOAD_openacc_async_test_all): Remove. (GOMP_OFFLOAD_openacc_async_wait): Remove. (GOMP_OFFLOAD_openacc_async_wait_async): Remove. (GOMP_OFFLOAD_openacc_async_wait_all): Remove. (GOMP_OFFLOAD_openacc_async_wait_all_async): Remove. (GOMP_OFFLOAD_openacc_async_set_async): Remove. (cuda_free_argmem): New function. (GOMP_OFFLOAD_openacc_async_exec): New plugin hook function. (GOMP_OFFLOAD_openacc_create_thread_data): Adjust code. (GOMP_OFFLOAD_openacc_cuda_get_stream): Adjust code. (GOMP_OFFLOAD_openacc_cuda_set_stream): Adjust code. (GOMP_OFFLOAD_openacc_async_construct): New plugin hook function. (GOMP_OFFLOAD_openacc_async_destruct): New plugin hook function. (GOMP_OFFLOAD_openacc_async_test): Remove and re-implement. (GOMP_OFFLOAD_openacc_async_synchronize): New plugin hook function. (GOMP_OFFLOAD_openacc_async_serialize): New plugin hook function. (GOMP_OFFLOAD_openacc_async_queue_callback): New plugin hook function. (cuda_callback_wrapper): New function. (cuda_memcpy_sanity_check): New function. (GOMP_OFFLOAD_host2dev): Remove and re-implement. (GOMP_OFFLOAD_dev2host): Remove and re-implement. (GOMP_OFFLOAD_openacc_async_host2dev): New plugin hook function. (GOMP_OFFLOAD_openacc_async_dev2host): New plugin hook function. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271128 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-08Address compiler diagnostics in libgomp.oacc-c-c++-common/pr87835.ctschwinge
source-gcc/libgomp/testsuite/libgomp.oacc-c-c++-common/pr87835.c: In function 'main': source-gcc/libgomp/testsuite/libgomp.oacc-c-c++-common/pr87835.c:45: warning: ignoring #pragma loop gang [-Wunknown-pragmas] 45 | #pragma loop gang | source-gcc/libgomp/testsuite/libgomp.oacc-c-c++-common/pr87835.c:19:7: warning: unused variable 'b' [-Wunused-variable] 19 | int b[n]; | ^ libgomp/ PR target/87835 * testsuite/libgomp.oacc-c-c++-common/pr87835.c: Update. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@271004 138bc75d-0d04-0410-961f-82ee72b054a4
2019-05-06Clean up libgomp GCC 5 legacy supporttschwinge
libgomp/ * oacc-parallel.c: Add comments to legacy entry points (GCC 5). git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@270901 138bc75d-0d04-0410-961f-82ee72b054a4
2019-03-27libgomp/ChangeLog:kevinb
* team.c (gomp_team_start): Initialize pool->threads[0]. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@269971 138bc75d-0d04-0410-961f-82ee72b054a4
2019-02-22[libgomp] In OpenACC testing, by default only build for the offload target ↵tschwinge
that we're actually going to test ... to avoid compilation overhead, and to keep simple '-foffload=[...]' handling in test cases. libgomp/ * testsuite/libgomp.oacc-c++/c++.exp: Specify "-foffload=$offload_target". * testsuite/libgomp.oacc-c/c.exp: Likewise. * testsuite/libgomp.oacc-fortran/fortran.exp: Likewise. * testsuite/lib/libgomp.exp (check_effective_target_openacc_nvidia_accel_configured): Remove, as (conceptually) merged into check_effective_target_openacc_nvidia_accel_selected. Adjust all users. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@269109 138bc75d-0d04-0410-961f-82ee72b054a4
2019-02-22[libgomp] In OpenACC testing, cycle though all offload targetstschwinge
... instead of through offload plugins. libgomp/ * plugin/configfrag.ac: Populate and AC_SUBST offload_targets. * testsuite/libgomp-test-support.exp.in: Adjust. * testsuite/lib/libgomp.exp: Likewise. Don't populate openacc_device_types_s. (offload_target_to_openacc_device_type): New proc. * testsuite/libgomp.oacc-c++/c++.exp: Adjust. * testsuite/libgomp.oacc-c/c.exp: Likewise. * testsuite/libgomp.oacc-fortran/fortran.exp: Likewise. * Makefile.in: Regenerate. * configure: Likewise. * testsuite/Makefile.in: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@269108 138bc75d-0d04-0410-961f-82ee72b054a4
2019-02-22[libgomp] Clarify difference between offload target, offload plugin, and ↵tschwinge
OpenACC device type libgomp/ * plugin/configfrag.ac: Populate and AC_SUBST offload_plugins instead of offload_targets, and AC_DEFINE_UNQUOTED OFFLOAD_PLUGINS instead of OFFLOAD_TARGETS. * target.c (gomp_target_init): Adjust. * testsuite/libgomp-test-support.exp.in: Likewise. * testsuite/lib/libgomp.exp: Likewise. Populate openacc_device_types_s instead of offload_targets_s_openacc. (check_effective_target_openacc_nvidia_accel_selected) (check_effective_target_openacc_host_selected): Adjust. * testsuite/libgomp.oacc-c++/c++.exp: Likewise. * testsuite/libgomp.oacc-c/c.exp: Likewise. * testsuite/libgomp.oacc-fortran/fortran.exp: Likewise. * Makefile.in: Regenerate. * config.h.in: Likewise. * configure: Likewise. * testsuite/Makefile.in: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@269107 138bc75d-0d04-0410-961f-82ee72b054a4
2019-02-22[libgomp] In OpenACC offloading testing, be more explicit in what is ↵tschwinge
supported, and what is not, or why not libgomp/ * testsuite/lib/libgomp.exp: Error out for unknown offload target. * testsuite/libgomp.oacc-c++/c++.exp: Likewise. Report if "offloading: supported, but hardware not accessible". * testsuite/libgomp.oacc-c/c.exp: Likewise. * testsuite/libgomp.oacc-fortran/fortran.exp: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@269106 138bc75d-0d04-0410-961f-82ee72b054a4
generated by cgit v1.2.3 (git 2.25.1) at 2024-06-10 11:02:15 +0000