Author: hahnfeld Date: Tue Oct 17 07:28:14 2017 New Revision: 316001 URL: http://llvm.org/viewvc/llvm-project?rev=316001&view=rev Log: [OpenMP] Implement omp_is_initial_device() as builtin
This allows to return the static value that we know at compile time. Differential Revision: https://reviews.llvm.org/D38968 Added: cfe/trunk/test/OpenMP/is_initial_device.c Modified: cfe/trunk/include/clang/Basic/Builtins.def cfe/trunk/include/clang/Basic/Builtins.h cfe/trunk/lib/AST/ExprConstant.cpp cfe/trunk/lib/Basic/Builtins.cpp Modified: cfe/trunk/include/clang/Basic/Builtins.def URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/Basic/Builtins.def?rev=316001&r1=316000&r2=316001&view=diff ============================================================================== --- cfe/trunk/include/clang/Basic/Builtins.def (original) +++ cfe/trunk/include/clang/Basic/Builtins.def Tue Oct 17 07:28:14 2017 @@ -1434,6 +1434,9 @@ LANGBUILTIN(__builtin_load_halff, "fhC*" BUILTIN(__builtin_os_log_format_buffer_size, "zcC*.", "p:0:nut") BUILTIN(__builtin_os_log_format, "v*v*cC*.", "p:0:nt") +// OpenMP 4.0 +LANGBUILTIN(omp_is_initial_device, "i", "nc", OMP_LANG) + // Builtins for XRay BUILTIN(__xray_customevent, "vcC*z", "") Modified: cfe/trunk/include/clang/Basic/Builtins.h URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/Basic/Builtins.h?rev=316001&r1=316000&r2=316001&view=diff ============================================================================== --- cfe/trunk/include/clang/Basic/Builtins.h (original) +++ cfe/trunk/include/clang/Basic/Builtins.h Tue Oct 17 07:28:14 2017 @@ -38,6 +38,7 @@ enum LanguageID { MS_LANG = 0x10, // builtin requires MS mode. OCLC20_LANG = 0x20, // builtin for OpenCL C 2.0 only. OCLC1X_LANG = 0x40, // builtin for OpenCL C 1.x only. + OMP_LANG = 0x80, // builtin requires OpenMP. ALL_LANGUAGES = C_LANG | CXX_LANG | OBJC_LANG, // builtin for all languages. ALL_GNU_LANGUAGES = ALL_LANGUAGES | GNU_LANG, // builtin requires GNU mode. ALL_MS_LANGUAGES = ALL_LANGUAGES | MS_LANG, // builtin requires MS mode. Modified: cfe/trunk/lib/AST/ExprConstant.cpp URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/AST/ExprConstant.cpp?rev=316001&r1=316000&r2=316001&view=diff ============================================================================== --- cfe/trunk/lib/AST/ExprConstant.cpp (original) +++ cfe/trunk/lib/AST/ExprConstant.cpp Tue Oct 17 07:28:14 2017 @@ -7929,6 +7929,9 @@ bool IntExprEvaluator::VisitBuiltinCallE return BuiltinOp == Builtin::BI__atomic_always_lock_free ? Success(0, E) : Error(E); } + case Builtin::BIomp_is_initial_device: + // We can decide statically which value the runtime would return if called. + return Success(Info.getLangOpts().OpenMPIsDevice ? 0 : 1, E); } } Modified: cfe/trunk/lib/Basic/Builtins.cpp URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Basic/Builtins.cpp?rev=316001&r1=316000&r2=316001&view=diff ============================================================================== --- cfe/trunk/lib/Basic/Builtins.cpp (original) +++ cfe/trunk/lib/Basic/Builtins.cpp Tue Oct 17 07:28:14 2017 @@ -75,8 +75,9 @@ bool Builtin::Context::builtinIsSupporte (BuiltinInfo.Langs & ALL_OCLC_LANGUAGES) == OCLC20_LANG; bool OclCUnsupported = !LangOpts.OpenCL && (BuiltinInfo.Langs & ALL_OCLC_LANGUAGES); + bool OpenMPUnsupported = !LangOpts.OpenMP && BuiltinInfo.Langs == OMP_LANG; return !BuiltinsUnsupported && !MathBuiltinsUnsupported && !OclCUnsupported && - !OclC1Unsupported && !OclC2Unsupported && + !OclC1Unsupported && !OclC2Unsupported && !OpenMPUnsupported && !GnuModeUnsupported && !MSModeUnsupported && !ObjCUnsupported; } Added: cfe/trunk/test/OpenMP/is_initial_device.c URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/test/OpenMP/is_initial_device.c?rev=316001&view=auto ============================================================================== --- cfe/trunk/test/OpenMP/is_initial_device.c (added) +++ cfe/trunk/test/OpenMP/is_initial_device.c Tue Oct 17 07:28:14 2017 @@ -0,0 +1,36 @@ +// REQUIRES: powerpc-registered-target + +// RUN: %clang_cc1 -verify -fopenmp -x c -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-unknown-unknown \ +// RUN: -emit-llvm-bc %s -o %t-ppc-host.bc +// RUN: %clang_cc1 -verify -fopenmp -x ir -triple powerpc64le-unknown-unknown -emit-llvm \ +// RUN: %t-ppc-host.bc -o - | FileCheck %s -check-prefixes HOST,OUTLINED +// RUN: %clang_cc1 -verify -fopenmp -x c -triple powerpc64le-unknown-unknown -emit-llvm -fopenmp-is-device \ +// RUN: %s -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s -check-prefixes DEVICE,OUTLINED + +// expected-no-diagnostics +int check() { + int host = omp_is_initial_device(); + int device; +#pragma omp target map(tofrom: device) + { + device = omp_is_initial_device(); + } + + return host + device; +} + +// The host should get a value of 1: +// HOST: define{{.*}} @check() +// HOST: [[HOST:%.*]] = alloca i32 +// HOST: store i32 1, i32* [[HOST]] + +// OUTLINED: define{{.*}} @{{.*}}omp_offloading{{.*}}(i32*{{.*}} [[DEVICE_ARGUMENT:%.*]]) +// OUTLINED: [[DEVICE_ADDR_STORAGE:%.*]] = alloca i32* +// OUTLINED: store i32* [[DEVICE_ARGUMENT]], i32** [[DEVICE_ADDR_STORAGE]] +// OUTLINED: [[DEVICE_ADDR:%.*]] = load i32*, i32** [[DEVICE_ADDR_STORAGE]] + +// The outlined function that is called as fallback also runs on the host: +// HOST: store i32 1, i32* [[DEVICE_ADDR]] + +// The device should get a value of 0: +// DEVICE: store i32 0, i32* [[DEVICE_ADDR]] _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
