gcc/libgomp/testsuite/libgomp.c-c++-common/target-is-initial-host-2.c

44 lines
1.5 KiB
C
Raw Permalink Normal View History

OpenMP: Constructors and destructors for "declare target" static aggregates This commit also compile-time expands (__builtin_)omp_is_initial_device for both Fortran and C/C++ (unless, -fno-builtin-omp_is_initial_device is used). But the main change is: This commit adds support for running constructors and destructors for static (file-scope) aggregates for C++ objects which are marked with "declare target" directives on OpenMP offload targets. Before this commit, space is allocated on the target for such aggregates, but nothing ever constructs them properly, so they end up zero-initialised. (See the new test static-aggr-constructor-destructor-3.C for a reason why running constructors on the target is preferable to e.g. constructing on the host and then copying the resulting object to the target.) 2024-08-07 Julian Brown <julian@codesourcery.com> Tobias Burnus <tobias@baylibre.com> gcc/ChangeLog: * builtins.def (DEF_GOMP_BUILTIN_COMPILER): Define DEF_GOMP_BUILTIN_COMPILER to handle the non-prefix version. * gimple-fold.cc (gimple_fold_builtin_omp_is_initial_device): New. (gimple_fold_builtin): Call it. * omp-builtins.def (BUILT_IN_OMP_IS_INITIAL_DEVICE): Define. * tree.cc (get_file_function_name): Support names for on-target constructor/destructor functions. gcc/cp/ * decl2.cc (tree-inline.h): Include. (static_init_fini_fns): Bump to four entries. Update comment. (start_objects, start_partial_init_fini_fn): Add 'omp_target' parameter. Support "declare target" decls. Update forward declaration. (emit_partial_init_fini_fn): Add 'host_fn' parameter. Return tree for the created function. Support "declare target". (OMP_SSDF_IDENTIFIER): New macro. (partition_vars_for_init_fini): Support partitioning "declare target" variables also. (generate_ctor_or_dtor_function): Add 'omp_target' parameter. Support "declare target" decls. (c_parse_final_cleanups): Support constructors/destructors on OpenMP offload targets. gcc/fortran/ChangeLog: * gfortran.h (gfc_option_t): Add disable_omp_is_initial_device. * lang.opt (fbuiltin-): Add. * options.cc (gfc_handle_option): Handle -fno-builtin-omp_is_initial_device. * f95-lang.cc (gfc_init_builtin_functions): Handle DEF_GOMP_BUILTIN_COMPILER. * trans-decl.cc (gfc_get_extern_function_decl): Add code to use DEF_GOMP_BUILTIN_COMPILER for 'omp_is_initial_device'. libgomp/ChangeLog: * testsuite/libgomp.c++/static-aggr-constructor-destructor-1.C: New test. * testsuite/libgomp.c++/static-aggr-constructor-destructor-2.C: New test. * testsuite/libgomp.c++/static-aggr-constructor-destructor-3.C: New test. * testsuite/libgomp.c-c++-common/target-is-initial-host.c: New test. * testsuite/libgomp.c-c++-common/target-is-initial-host-2.c: New test. * testsuite/libgomp.fortran/target-is-initial-host.f: New test. * testsuite/libgomp.fortran/target-is-initial-host.f90: New test. * testsuite/libgomp.fortran/target-is-initial-host-2.f90: New test. Co-authored-by: Tobias Burnus <tobias@baylibre.com>
2024-08-07 19:31:19 +02:00
/* { dg-do run } */
/* { dg-additional-options "-fno-builtin-omp_is_initial_device" } */
/* Check whether 'omp_is_initial_device()' is NOT compile-time optimized. */
/* { dg-additional-options "-fdump-tree-gimple -fdump-tree-optimized" } */
/* { dg-additional-options -foffload-options=-fdump-tree-optimized { target { offload_target_nvptx || offload_target_amdgcn } } } */
OpenMP: Constructors and destructors for "declare target" static aggregates This commit also compile-time expands (__builtin_)omp_is_initial_device for both Fortran and C/C++ (unless, -fno-builtin-omp_is_initial_device is used). But the main change is: This commit adds support for running constructors and destructors for static (file-scope) aggregates for C++ objects which are marked with "declare target" directives on OpenMP offload targets. Before this commit, space is allocated on the target for such aggregates, but nothing ever constructs them properly, so they end up zero-initialised. (See the new test static-aggr-constructor-destructor-3.C for a reason why running constructors on the target is preferable to e.g. constructing on the host and then copying the resulting object to the target.) 2024-08-07 Julian Brown <julian@codesourcery.com> Tobias Burnus <tobias@baylibre.com> gcc/ChangeLog: * builtins.def (DEF_GOMP_BUILTIN_COMPILER): Define DEF_GOMP_BUILTIN_COMPILER to handle the non-prefix version. * gimple-fold.cc (gimple_fold_builtin_omp_is_initial_device): New. (gimple_fold_builtin): Call it. * omp-builtins.def (BUILT_IN_OMP_IS_INITIAL_DEVICE): Define. * tree.cc (get_file_function_name): Support names for on-target constructor/destructor functions. gcc/cp/ * decl2.cc (tree-inline.h): Include. (static_init_fini_fns): Bump to four entries. Update comment. (start_objects, start_partial_init_fini_fn): Add 'omp_target' parameter. Support "declare target" decls. Update forward declaration. (emit_partial_init_fini_fn): Add 'host_fn' parameter. Return tree for the created function. Support "declare target". (OMP_SSDF_IDENTIFIER): New macro. (partition_vars_for_init_fini): Support partitioning "declare target" variables also. (generate_ctor_or_dtor_function): Add 'omp_target' parameter. Support "declare target" decls. (c_parse_final_cleanups): Support constructors/destructors on OpenMP offload targets. gcc/fortran/ChangeLog: * gfortran.h (gfc_option_t): Add disable_omp_is_initial_device. * lang.opt (fbuiltin-): Add. * options.cc (gfc_handle_option): Handle -fno-builtin-omp_is_initial_device. * f95-lang.cc (gfc_init_builtin_functions): Handle DEF_GOMP_BUILTIN_COMPILER. * trans-decl.cc (gfc_get_extern_function_decl): Add code to use DEF_GOMP_BUILTIN_COMPILER for 'omp_is_initial_device'. libgomp/ChangeLog: * testsuite/libgomp.c++/static-aggr-constructor-destructor-1.C: New test. * testsuite/libgomp.c++/static-aggr-constructor-destructor-2.C: New test. * testsuite/libgomp.c++/static-aggr-constructor-destructor-3.C: New test. * testsuite/libgomp.c-c++-common/target-is-initial-host.c: New test. * testsuite/libgomp.c-c++-common/target-is-initial-host-2.c: New test. * testsuite/libgomp.fortran/target-is-initial-host.f: New test. * testsuite/libgomp.fortran/target-is-initial-host.f90: New test. * testsuite/libgomp.fortran/target-is-initial-host-2.f90: New test. Co-authored-by: Tobias Burnus <tobias@baylibre.com>
2024-08-07 19:31:19 +02:00
/* { dg-final { scan-tree-dump-times "omp_is_initial_device" 1 "gimple" } } */
/* { dg-final { scan-tree-dump-times "omp_is_initial_device" 1 "optimized" } } */
/* { dg-final { only_for_offload_target amdgcn-amdhsa scan-offload-tree-dump-times "omp_is_initial_device" 1 "optimized" { target offload_target_amdgcn } } } */
/* { dg-final { only_for_offload_target nvptx-none scan-offload-tree-dump-times "omp_is_initial_device" 1 "optimized" { target offload_target_nvptx } } } */
#include <omp.h>
int
main ()
{
int is_initial, dev_num, initial;
initial = omp_get_initial_device();
for (int dev = omp_initial_device; dev <= omp_get_num_devices(); dev++)
{
is_initial = dev_num = 99;
#pragma omp target map(from: is_initial, dev_num) device(dev)
{
is_initial = omp_is_initial_device ();
dev_num = omp_get_device_num ();
}
if (dev == omp_initial_device || dev == initial)
{
if (dev_num != initial || is_initial != 1)
__builtin_abort ();
}
else
{
if (dev_num != dev || is_initial != 0)
__builtin_abort ();
}
}
}