Commit Graph

958 Commits

Author SHA1 Message Date
Tobias Burnus ffbdd78a4a Fortran] OpenACC – libgomp/testsuite – use 'stop' and 'dg-do run'
* testsuite/libgomp.oacc-fortran/abort-1.f90: Add 'dg-do run'.
        * testsuite/libgomp.oacc-fortran/abort-2.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/acc_on_device-1-1.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/acc_on_device-1-2.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/acc_on_device-1-3.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/lib-1.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/common-block-1.f90:
        Use 'stop' not abort().
        * testsuite/libgomp.oacc-fortran/common-block-2.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/common-block-3.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/data-1.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/data-2.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/data-5.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/dummy-array.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/gemm-2.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/gemm.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/host_data-2.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/host_data-3.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/host_data-4.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-collapse-3.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-collapse-4.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-independent.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-loop-1.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-map-1.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-parallel-loop-data-enter-exit.f95:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-loop-gang-1.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-loop-gang-2.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-loop-gang-3.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-loop-gang-6.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-vector-1.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-vector-2.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-worker-1.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-worker-2.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-worker-3.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-worker-4.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-worker-5.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-worker-6.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-private-vars-worker-7.f90:
        Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-reduction-1.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/lib-12.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/lib-13.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/lib-14.f90: Ditto.
        * testsuite/libgomp.oacc-fortran/kernels-acc-loop-reduction-2.f90:
        Likewise and also add 'dg-do run'.
        * testsuite/libgomp.oacc-fortran/kernels-acc-loop-reduction.f90:
        Ditto.

From-SVN: r277503
2019-10-28 08:39:26 +01:00
Cesar Philippidis ec0846134e [Fortran] OpenACC – permit common blocks in some clauses
2019-10-25  Cesar Philippidis <cesar@codesourcery.com>
	    Tobias Burnus  <tobias@codesourcery.com>

	gcc/fortran/
	* openmp.c (gfc_match_omp_map_clause): Add and pass allow_commons
	argument.
	(gfc_match_omp_clauses): Update calls to permit common blocks for
	OpenACC's copy/copyin/copyout, create/delete, host,
	pcopy/pcopy_in/pcopy_out, present_or_copy, present_or_copy_in,
	present_or_copy_out, present_or_create and self.

	gcc/
	* gimplify.c (oacc_default_clause): Privatize fortran common blocks.
	(omp_notice_variable): Defer the expansion of DECL_VALUE_EXPR for
	common block decls.
    
	gcc/testsuite/
	* gfortran.dg/goacc/common-block-1.f90: New test.
	* gfortran.dg/goacc/common-block-2.f90: New test.
	* gfortran.dg/goacc/common-block-3.f90: New test.
    
	libgomp/
	* testsuite/libgomp.oacc-fortran/common-block-1.f90: New test.
	* testsuite/libgomp.oacc-fortran/common-block-2.f90: New test.
	* testsuite/libgomp.oacc-fortran/common-block-3.f90: New test.

Reviewed-by: Thomas Schwinge <thomas@codesourcery.com>

Co-Authored-By: Tobias Burnus <tobias@codesourcery.com>

From-SVN: r277451
2019-10-25 16:28:40 +02:00
Jakub Jelinek 77ef339456 re PR libgomp/92081 (FAIL: libgomp.fortran/target-simd.f90 execution test)
PR libgomp/92081
	* testsuite/libgomp.fortran/target-simd.f90: Iterate from 1 rather
	than 0.

From-SVN: r276956
2019-10-14 10:48:42 +02:00
Tobias Burnus 08c14aaaab [OpenMP,Fortran] Fix several OpenMP use_device_addr/map/update errors
gcc/fortran/
	* f95-lang.c (LANG_HOOKS_OMP_IS_ALLOCATABLE_OR_PTR): Re-define to
	gfc_omp_is_allocatable_or_ptr.
	* trans-decl.c (create_function_arglist): Set GFC_DECL_OPTIONAL_ARGUMENT
	only if not passed by value.
	* trans-openmp.c (gfc_omp_is_allocatable_or_ptr): New.
	(gfc_trans_omp_clauses): For MAP, handle (present) optional arguments;
	for target update, handle allocatable/pointer scalars.
	* trans.h (gfc_omp_is_allocatable_or_ptr): Declare.

	gcc/
	* langhooks-def.h (LANG_HOOKS_OMP_IS_ALLOCATABLE_OR_PTR): Define.
	(LANG_HOOKS_DECLS): Add it.
	* langhooks.h (lang_hooks_for_decls): Add omp_is_allocatable_or_ptr;
	update comment for omp_is_optional_argument.
	* omp-general.c (omp_is_allocatable_or_ptr): New.
	* omp-general.h (omp_is_allocatable_or_ptr): Declare.
	* omp-low.c (scan_sharing_clauses, lower_omp_target): Handle
	Fortran's optional arguments and allocatable/pointer scalars
	with use_device_addr.

	libgomp/
	* testsuite/libgomp.fortran/use_device_addr-1.f90: New.
	* testsuite/libgomp.fortran/use_device_addr-2.f90: New.

From-SVN: r276875
2019-10-11 11:17:49 +02:00
Thomas Schwinge 6bbead0c5a [PR92036] Add 'libgomp.oacc-c-c++-common/data-firstprivate-1.c'
libgomp/
	PR middle-end/92036
	* testsuite/libgomp.oacc-c-c++-common/data-firstprivate-1.c: New
	file.

From-SVN: r276757
2019-10-09 13:31:14 +02:00
Tobias Burnus eba3dee8eb re PR testsuite/91884 (libgomp testsuite: (not) using a specific driver for C++, Fortran)
2019-10-09  Tobias Burnus  <tobias@codesourcery.com>

        PR testsuite/91884
        * testsuite/libgomp.fortran/fortran.exp: Conditionally
        add -lquadmath.
        * testsuite/libgomp.oacc-fortran/fortran.exp: Ditto.

From-SVN: r276754
2019-10-09 10:37:44 +02:00
Jakub Jelinek 7d48e14fc6 re PR libgomp/92028 (OpenACC 'host_data' execution test regressions with nvptx offloading)
PR libgomp/92028
	* target.c (gomp_map_vars_internal): Readd the previous
	GOMP_MAP_USE_DEVICE_PTR handling code in the first loop,
	though do that just in the !not_found_cnt case.

From-SVN: r276753
2019-10-09 09:33:02 +02:00
Tobias Burnus 65b67cf390 Fortran - fix OpenMP 'target simd'
gcc/fortran/
	* parse.c (parse_executable): Add missing ST_OMP_TARGET_SIMD.

	libgomp/
	* testsuite/libgomp.fortran/target-simd.f90: New.

From-SVN: r276698
2019-10-08 14:30:44 +02:00
Julian Brown 6c7e076b74 Libgomp magic offset value self-documentation
2019-10-02  Julian Brown  <julian@codesourcery.com>
	    Cesar Philippidis  <cesar@codesourcery.com>

	libgomp/
	* libgomp.h (OFFSET_INLINED, OFFSET_POINTER, OFFSET_STRUCT): Define.
	* target.c (FIELD_TGT_EMPTY): Define.
	(gomp_map_val): Use OFFSET_* macros instead of magic constants.  Write
	as switch instead of list of ifs.
	(gomp_map_vars_internal): Use OFFSET_* and FIELD_TGT_EMPTY macros.

Co-Authored-By: Cesar Philippidis <cesar@codesourcery.com>

From-SVN: r276519
2019-10-03 17:48:36 +00:00
Andreas Tobler d61bff850d Testsuite, remove alloca header
2019-10-02  Andreas Tobler  <andreast@gcc.gnu.org>

	* testsuite/libgomp.oacc-c-c++-common/loop-default.h: Remove alloca.h
	include. Replace alloca () with __builtin_alloca ().
	* testsuite/libgomp.oacc-c-c++-common/loop-dim-default.c: Likewise.

Reviewed-by: Thomas Schwinge <thomas@codesourcery.com>

From-SVN: r276479
2019-10-02 21:05:35 +02:00
Tobias Burnus 67c259509c Fix omp target issue with Fortran optional arguments
gcc/
        * omp-low.c (lower_omp_target): Dereference optional argument
        to work with the right pointer.

        gcc/testsuite/
        * libgomp/testsuite/libgomp.fortran/use_device_ptr-optional-1.f90: New.

From-SVN: r276445
2019-10-02 12:33:42 +02:00
Jakub Jelinek 810f316dd6 configure.ac: Remove GCC_HEADER_STDINT(gstdint.h).
* configure.ac: Remove GCC_HEADER_STDINT(gstdint.h).
	* libgomp.h: Include <stdint.h> instead of "gstdint.h".
	* oacc-parallel.c: Don't include "libgomp_g.h".
	* plugin/plugin-hsa.c: Include <stdint.h> instead of "gstdint.h".
	* plugin/plugin-nvptx.c: Don't include "gstdint.h".
	* aclocal.m4: Regenerated.
	* config.h.in: Regenerated.
	* configure: Regenerated.
	* Makefile.in: Regenerated.

From-SVN: r276389
2019-10-01 09:51:46 +02:00
Kwok Cheung Yeung d7f9ee981f libgomp_g.h: Include stdint.h instead of gstdint.h.
2019-09-30  Kwok Cheung Yeung  <kcy@codesourcery.com>

        * libgomp_g.h: Include stdint.h instead of gstdint.h.

From-SVN: r276301
2019-09-30 16:16:34 +02:00
Maciej W. Rozycki e9085da528 Regenerate `configure' scripts for `uclinuxfdpiceabi' libtool.m4 update
A change made with r275564 ("[ARM/FDPIC v6 02/24] [ARM] FDPIC: Handle 
arm*-*-uclinuxfdpiceabi in configure scripts") to libtool.m4 has not 
regenerated all the `configure' scripts affected.  Fix it.

	gcc/
	* configure: Regenerate.

	libatomic/
	* configure: Regenerate.

	libbacktrace/
	* configure: Regenerate.

	libcc1/
	* configure: Regenerate.

	libffi/
	* configure: Regenerate.

	libgfortran/
	* configure: Regenerate.

	libgomp/
	* configure: Regenerate.

	libhsail-rt/
	* configure: Regenerate.

	libitm/
	* configure: Regenerate.

	libobjc/
	* configure: Regenerate.

	liboffloadmic/
	* configure: Regenerate.

	libphobos/
	* configure: Regenerate.

	libquadmath/
	* configure: Regenerate.

	libsanitizer/
	* configure: Regenerate.

	libssp/
	* configure: Regenerate.

	libstdc++-v3/
	* configure: Regenerate.

	libvtv/
	* configure: Regenerate.

	lto-plugin/
	* configure: Regenerate.

	zlib/
	* configure: Regenerate.

From-SVN: r276213
2019-09-27 21:24:42 +00:00
Tobias Burnus c28712beb4 libgomp plugin - init string
libgomp/
2019-09-13  Tobias Burnus  <tobias@codesourcery.com>

        * plugin/plugin-hsa.c (hsa_warn, hsa_fatal, hsa_error): Ensure
        string is initialized.

From-SVN: r275703
2019-09-13 20:14:02 +02:00
Florian Weimer 3c086f8dec Fix GCC_LINUX_FUTEX to work with C99 compilers
Without this change, libstdc++ is built without futex symbols if GCC
rejects implicit function declarations by default.

From-SVN: r275454
2019-09-06 12:27:36 +02:00
Chung-Lin Tang c6c2d1bc9b re PR other/79543 (Inappropriate "ld --version" checking)
2019-09-03  Chung-Lin Tang <cltang@codesourcery.com>

	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.

From-SVN: r275341
2019-09-03 14:10:26 +00:00
Jakub Jelinek e62506f362 re PR libgomp/91530 (Several libgomp.*/scan-* tests FAIL without avx_runtime)
PR libgomp/91530
	* 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.

From-SVN: r274985
2019-08-28 12:13:21 +02:00
Jakub Jelinek 5cb72d83bb re PR libgomp/91530 (Several libgomp.*/scan-* tests FAIL without avx_runtime)
PR libgomp/91530
	* 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.

From-SVN: r274984
2019-08-28 12:12:11 +02:00
Jakub Jelinek 0ad7981cb4 re PR libgomp/91530 (Several libgomp.*/scan-* tests FAIL without avx_runtime)
PR libgomp/91530
	* 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.

From-SVN: r274947
2019-08-27 12:45:55 +02:00
Thomas Koenig 1e67491a0d re PR libgomp/91473 (Test case libgomp.fortran/appendix-a/a.28.5.f90 is invalid)
2019-08-17  Thomas Koenig  <tkoenig@gcc.gnu.org>

    PR fortran/91473
    * testsuite/libgomp.fortran/appendix-a/a.28.5.f90: Add
    -std=legacy so invalid code in the test case is accepted.

From-SVN: r274602
2019-08-17 11:57:25 +00:00
Thomas Koenig 393fdeb1e4 re PR fortran/91424 (Extend warnings about DO loops)
2019-08-12  Thomas Koenig  <tkoenig@gcc.gnu.org>

	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.

From-SVN: r274320
2019-08-12 20:21:37 +00:00
Jakub Jelinek 8860d2706d gimplify.c (omp_add_variable): Use GOVD_PRIVATE | GOVD_EXPLICIT for VLA helper variables on target data even if...
* gimplify.c (omp_add_variable): Use GOVD_PRIVATE | GOVD_EXPLICIT
	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.

From-SVN: r274206
2019-08-08 08:39:02 +02:00
Jakub Jelinek 398e3feb8a tree-core.h (enum omp_clause_code): Adjust OMP_CLAUSE_USE_DEVICE_PTR OpenMP description.
* tree-core.h (enum omp_clause_code): Adjust OMP_CLAUSE_USE_DEVICE_PTR
	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.

From-SVN: r274159
2019-08-07 09:27:10 +02:00
Jakub Jelinek d81ab49d05 tree.h (OMP_CLAUSE_LASTPRIVATE_TASKLOOP_IV): Rename to ...
* tree.h (OMP_CLAUSE_LASTPRIVATE_TASKLOOP_IV): Rename to ...
	(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.

From-SVN: r274138
2019-08-06 09:26:32 +02:00
Jakub Jelinek c3ac76aa40 re PR middle-end/91301 (ICE in omp_add_variable on random access iterator distribute parallel for private (iterator))
PR middle-end/91301
	* 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.

From-SVN: r273922
2019-07-31 09:49:56 +02:00
Steven G. Kargl 8dc63166e0 arith.c (gfc_convert_integer, [...]): Move to ...
2019-07-23  Steven G. Kargl  <kargl@gcc.gnu.org>

	* 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.

From-SVN: r273747
2019-07-23 21:43:21 +00:00
Jakub Jelinek 554a530ff8 tree.def (OMP_LOOP): New tree code.
* tree.def (OMP_LOOP): New tree code.
	* 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.

From-SVN: r273621
2019-07-20 13:21:42 +02:00
Jakub Jelinek 6cda84b509 tree-vect-stmts.c (scan_operand_equal_p): Look through MEM_REF with SSA_NAME address of POINTER_PLUS_EXPR.
* tree-vect-stmts.c (scan_operand_equal_p): Look through MEM_REF
	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.

From-SVN: r273249
2019-07-09 00:11:59 +02:00
Jakub Jelinek 6f67abcdb0 omp-low.c (lower_rec_input_clauses): For lastprivate clauses in ctx->for_simd_scan_phase simd copy the outer var to...
* omp-low.c (lower_rec_input_clauses): For lastprivate clauses in
	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.

From-SVN: r273169
2019-07-06 23:58:01 +02:00
Jakub Jelinek 1f52d1a8b5 omp-low.c (struct omp_context): Add for_simd_scan_phase member.
* omp-low.c (struct omp_context): Add for_simd_scan_phase member.
	(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.

From-SVN: r273157
2019-07-06 09:53:48 +02:00
Jakub Jelinek 2f03073ff2 omp-expand.c (expand_omp_for_static_nochunk): Don't emit GOMP_loop_start at the start of second worksharing loop in a scan.
* omp-expand.c (expand_omp_for_static_nochunk): Don't emit
	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.

From-SVN: r273095
2019-07-04 23:40:56 +02:00
Jakub Jelinek 2f6bb511d1 tree-core.h (enum omp_clause_code): Add OMP_CLAUSE__SCANTEMP_ clause.
* tree-core.h (enum omp_clause_code): Add OMP_CLAUSE__SCANTEMP_
	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.

From-SVN: r272958
2019-07-03 07:03:58 +02:00
Thomas Schwinge 85fca03a09 Test cases to verify OpenACC 'firstprivate' mappings
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.

From-SVN: r272451
2019-06-19 00:15:03 +02:00
Thomas Schwinge bd194a51d4 Add missing results check in 'libgomp.fortran/allocatable3.f90'
libgomp/
	* testsuite/libgomp.fortran/allocatable3.f90: Add missing results
	check.

From-SVN: r272449
2019-06-19 00:14:43 +02:00
Cesar Philippidis 6652161ef3 Add 'libgomp.oacc-fortran/allocatable-array-1.f90'
libgomp/
	* testsuite/libgomp.oacc-fortran/allocatable-array-1.f90: New
	file.

From-SVN: r272448
2019-06-19 00:14:34 +02:00
Thomas Schwinge 4017da8d1c [PR90743] Fortran 'allocatable' with OpenACC data/OpenMP 'target' 'map' clauses
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.

From-SVN: r272447
2019-06-19 00:14:24 +02:00
Thomas Schwinge 6f7c1f6502 [PR90861] Document status quo for OpenACC 'declare' not cleaning up for VLAs
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.

From-SVN: r272446
2019-06-19 00:14:14 +02:00
Thomas Schwinge 3a37d6f68c [PR90862] OpenACC 'declare' ICE when nested inside another construct
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.

From-SVN: r272444
2019-06-19 00:13:54 +02:00
Tom de Vries f45ce17d98 [openacc, parloops] Fix SIGSEGV in oacc_entry_exit_ok_1
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.

From-SVN: r272338
2019-06-16 07:47:15 +00:00
Tom de Vries 00908992f2 [nvptx, libgomp] Update pr85381-{2,4}.c test-cases
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.

From-SVN: r272324
2019-06-15 12:01:30 +00:00
Jakub Jelinek 211b7533bf re PR middle-end/90779 (Fortran array initialization in offload regions)
PR middle-end/90779
	* 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.

From-SVN: r272322
2019-06-15 09:09:04 +02:00
Tom de Vries 120a01d160 [openacc] Disable pass_thread_jumps for IFN_UNIQUE
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.

From-SVN: r272321
2019-06-15 07:06:19 +00:00
Feng Xue c29c92c789 PR tree-optimization/89713 - Assume loop with an exit is finite
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.

From-SVN: r272234
2019-06-13 04:17:42 +00:00
Jakub Jelinek ce9c4ec3c5 re PR target/90811 ([nvptx] ptxas error on OpenMP offloaded code)
PR target/90811
	* config/nvptx/nvptx.c (nvptx_output_softstack_switch): Use and.b%d
	instead of and.u%d.

	* testsuite/libgomp.c/pr90811.c: New test.

From-SVN: r272161
2019-06-11 18:40:10 +02:00
Jakub Jelinek 28b3a77ca0 omp-low.c (lower_rec_input_clauses): For lastprivate conditional references...
* omp-low.c (lower_rec_input_clauses): For lastprivate conditional
	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.

From-SVN: r271948
2019-06-05 09:37:40 +02:00
Jakub Jelinek 7855700e63 gimplify.c (gimplify_scan_omp_clauses): Don't sorry_at on lastprivate conditional on combined for simd.
* gimplify.c (gimplify_scan_omp_clauses): Don't sorry_at on lastprivate
	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.

From-SVN: r271907
2019-06-04 14:49:03 +02:00
Rainer Orth a7155c2e0b Generalize getconf _NPROCESSORS_ONLN
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.

From-SVN: r271769
2019-05-30 09:06:48 +00:00
Jakub Jelinek 7e47198b80 gimplify.c (struct gimplify_omp_ctx): Add clauses member.
* gimplify.c (struct gimplify_omp_ctx): Add clauses member.
	(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.

From-SVN: r271733
2019-05-29 09:51:43 +02:00
Jakub Jelinek 8e7757ba17 gimplify.c (gimplify_scan_omp_clauses): Allow lastprivate conditional on sections construct.
* gimplify.c (gimplify_scan_omp_clauses): Allow lastprivate conditional
	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.

From-SVN: r271673
2019-05-27 23:33:37 +02:00