ObjC,
ObjCXX,
OpenCL,
+ OpenCLCXX,
CUDA,
RenderScript,
HIP,
TYPE("cpp-output", PP_C, INVALID, "i", phases::Compile, phases::Backend, phases::Assemble, phases::Link)
TYPE("c", C, PP_C, "c", phases::Preprocess, phases::Compile, phases::Backend, phases::Assemble, phases::Link)
TYPE("cl", CL, PP_C, "cl", phases::Preprocess, phases::Compile, phases::Backend, phases::Assemble, phases::Link)
+TYPE("clcpp", CLCXX, PP_CXX, "clcpp", phases::Preprocess, phases::Compile, phases::Backend, phases::Assemble, phases::Link)
TYPE("cuda-cpp-output", PP_CUDA, INVALID, "cui", phases::Compile, phases::Backend, phases::Assemble, phases::Link)
TYPE("cuda", CUDA, PP_CUDA, "cu", phases::Preprocess, phases::Compile, phases::Backend, phases::Assemble, phases::Link)
TYPE("cuda", CUDA_DEVICE, PP_CUDA, "cu", phases::Preprocess, phases::Compile, phases::Backend, phases::Assemble, phases::Link)
case TY_Asm:
case TY_C: case TY_PP_C:
- case TY_CL:
+ case TY_CL: case TY_CLCXX:
case TY_CUDA: case TY_PP_CUDA:
case TY_CUDA_DEVICE:
case TY_HIP:
}
}
-bool types::isOpenCL(ID Id) { return Id == TY_CL; }
+bool types::isOpenCL(ID Id) { return Id == TY_CL || Id == TY_CLCXX; }
bool types::isCXX(ID Id) {
switch (Id) {
.Case("cc", TY_CXX)
.Case("CC", TY_CXX)
.Case("cl", TY_CL)
+ .Case("clcpp", TY_CLCXX)
.Case("cp", TY_CXX)
.Case("cu", TY_CUDA)
.Case("hh", TY_CXXHeader)
case types::TY_ObjCXX:
return types::TY_ObjCXXHeader;
case types::TY_CL:
+ case types::TY_CLCXX:
return types::TY_CLHeader;
}
}
static unsigned getOptimizationLevel(ArgList &Args, InputKind IK,
DiagnosticsEngine &Diags) {
unsigned DefaultOpt = llvm::CodeGenOpt::None;
- if (IK.getLanguage() == Language::OpenCL && !Args.hasArg(OPT_cl_opt_disable))
+ if ((IK.getLanguage() == Language::OpenCL ||
+ IK.getLanguage() == Language::OpenCLCXX) &&
+ !Args.hasArg(OPT_cl_opt_disable))
DefaultOpt = llvm::CodeGenOpt::Default;
if (Arg *A = Args.getLastArg(options::OPT_O_Group)) {
case Language::OpenCL:
Lang = "cl";
break;
+ case Language::OpenCLCXX:
+ Lang = "clcpp";
+ break;
case Language::CUDA:
Lang = "cuda";
break;
DashX = llvm::StringSwitch<InputKind>(XValue)
.Case("c", Language::C)
.Case("cl", Language::OpenCL)
+ .Case("clcpp", Language::OpenCLCXX)
.Case("cuda", Language::CUDA)
.Case("hip", Language::HIP)
.Case("c++", Language::CXX)
case Language::OpenCL:
LangStd = LangStandard::lang_opencl10;
break;
+ case Language::OpenCLCXX:
+ LangStd = LangStandard::lang_openclcpp;
+ break;
case Language::CUDA:
LangStd = LangStandard::lang_cuda;
break;
return S.getLanguage() == Language::C;
case Language::OpenCL:
- return S.getLanguage() == Language::OpenCL;
+ return S.getLanguage() == Language::OpenCL ||
+ S.getLanguage() == Language::OpenCLCXX;
+
+ case Language::OpenCLCXX:
+ return S.getLanguage() == Language::OpenCLCXX;
case Language::CXX:
case Language::ObjCXX:
return "Objective-C++";
case Language::OpenCL:
return "OpenCL";
+ case Language::OpenCLCXX:
+ return "C++ for OpenCL";
case Language::CUDA:
return "CUDA";
case Language::RenderScript:
case Language::ObjC:
case Language::ObjCXX:
case Language::OpenCL:
+ case Language::OpenCLCXX:
case Language::CUDA:
case Language::HIP:
break;
.Case("cppm", Language::CXX)
.Case("iim", InputKind(Language::CXX).getPreprocessed())
.Case("cl", Language::OpenCL)
+ .Case("clcpp", Language::OpenCLCXX)
.Cases("cu", "cuh", Language::CUDA)
.Case("hip", Language::HIP)
.Cases("ll", "bc", Language::LLVM_IR)
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -O0 -emit-llvm -o - | FileCheck %s -check-prefixes=COMMON,PTR
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -O0 -emit-llvm -o - -DREF | FileCheck %s -check-prefixes=COMMON,REF
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -O0 -emit-llvm -o - | FileCheck %s -check-prefixes=COMMON,PTR
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -O0 -emit-llvm -o - -DREF | FileCheck %s -check-prefixes=COMMON,REF
#ifdef REF
#define PTR &
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -O0 -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -O0 -emit-llvm -o - | FileCheck %s
class P {
public:
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -emit-llvm -O0 -o - | FileCheck %s
void bar(__generic volatile unsigned int* ptr)
{
-// RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir -emit-llvm -O0 -o - | FileCheck %s
struct B {
int mb;
-// RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir -emit-llvm -O0 -o - | FileCheck %s
typedef __SIZE_TYPE__ size_t;
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -emit-llvm -pedantic -verify -O0 -o - -DDECL | FileCheck %s --check-prefixes="COMMON,EXPL"
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -emit-llvm -pedantic -verify -O0 -o - -DDECL -DUSE_DEFLT | FileCheck %s --check-prefixes="COMMON,IMPL"
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -emit-llvm -pedantic -verify -O0 -o - | FileCheck %s --check-prefixes="COMMON,IMPL"
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -emit-llvm -pedantic -verify -O0 -o - -DDECL | FileCheck %s --check-prefixes="COMMON,EXPL"
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -emit-llvm -pedantic -verify -O0 -o - -DDECL -DUSE_DEFLT | FileCheck %s --check-prefixes="COMMON,IMPL"
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -emit-llvm -pedantic -verify -O0 -o - | FileCheck %s --check-prefixes="COMMON,IMPL"
// expected-no-diagnostics
// Test that the 'this' pointer is in the __generic address space.
-//RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+//RUN: %clang_cc1 %s -triple spir -emit-llvm -O0 -o - | FileCheck %s
enum E {
a,
-//RUN: %clang_cc1 %s -cl-std=clc++ -triple spir -emit-llvm -o - -O0 | FileCheck %s
+//RUN: %clang_cc1 %s -triple spir -emit-llvm -o - -O0 | FileCheck %s
typedef short short2 __attribute__((ext_vector_type(2)));
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=CLC++ -emit-llvm -O0 -o - | FileCheck %s --check-prefix=CHECK-DEFINITIONS
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -emit-llvm -O0 -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -emit-llvm -O0 -o - | FileCheck %s --check-prefix=CHECK-DEFINITIONS
// This test ensures the proper address spaces and address space cast are used
// for constructors, member functions and destructors.
-//RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+//RUN: %clang_cc1 %s -triple spir -emit-llvm -O0 -o - | FileCheck %s
//CHECK-LABEL: define{{.*}} spir_func void @_Z3barPU3AS1i
void bar(global int *gl) {
-//RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+//RUN: %clang_cc1 %s -triple spir -emit-llvm -O0 -o - | FileCheck %s
struct S {
~S(){};
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -O0 -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -O0 -emit-llvm -o - | FileCheck %s
typedef int int2 __attribute__((ext_vector_type(2)));
typedef int int4 __attribute__((ext_vector_type(4)));
-// RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir -emit-llvm -O0 -o - | FileCheck %s
struct S {
S() {}
-// RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir -emit-llvm -O0 -o - | FileCheck %s
// Test that we don't initialize local address space objects.
//CHECK: @_ZZ4testE1i = internal addrspace(3) global i32 undef
-//RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -emit-llvm -O0 -o - | FileCheck %s
+//RUN: %clang_cc1 %s -triple spir-unknown-unknown -emit-llvm -O0 -o - | FileCheck %s
struct C {
void foo() __local;
-// RUN: %clang_cc1 -cl-std=clc++ %s -emit-llvm -o - -O0 -triple spir-unknown-unknown | FileCheck %s
+// RUN: %clang_cc1 %s -emit-llvm -o - -O0 -triple spir-unknown-unknown | FileCheck %s
template <typename T>
struct S{
--- /dev/null
+// RUN: %clang %s -Xclang -verify -fsyntax-only
+// RUN: %clang %s -cl-std=clc++ -Xclang -verify -fsyntax-only
+// RUN: %clang %s -cl-std=cl2.0 -Xclang -verify -fsyntax-only
+// RUN: %clang %s -### 2>&1 | FileCheck %s
+
+// CHECK: "-x" "clcpp"
+
+#ifdef __OPENCL_CPP_VERSION__
+//expected-no-diagnostics
+#endif
+
+kernel void k(){
+ auto a = get_local_id(1);
+#ifndef __OPENCL_CPP_VERSION__
+//expected-error@-2{{OpenCL C version 2.0 does not support the 'auto' storage class specifier}}
+//expected-warning@-3{{type specifier missing, defaults to 'int'}}
+#endif
+}
config.suffixes = ['.c', '.cpp', '.h', '.m', '.mm', '.S', '.s', '.f90', '.F90', '.f95',
- '.cu', '.rs', '.cl', '.hip']
+ '.cu', '.rs', '.cl', '.clcpp', '.hip']
config.substitutions = list(config.substitutions)
config.substitutions.insert(0,
('%clang_cc1',
-//RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -ast-dump -verify | FileCheck %s
+//RUN: %clang_cc1 %s -pedantic -ast-dump -verify | FileCheck %s
void nester_ptr() {
local int * * locgen;
-// RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -verify
+// RUN: %clang_cc1 %s -pedantic -verify
namespace PointerRvalues {
-//RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -ast-dump -verify | FileCheck %s
+//RUN: %clang_cc1 %s -pedantic -ast-dump -verify | FileCheck %s
//expected-no-diagnostics
-//RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -ast-dump -verify | FileCheck %s
-//RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -ast-dump -verify -triple i386-windows | FileCheck %s
+//RUN: %clang_cc1 %s -pedantic -ast-dump -verify | FileCheck %s
+//RUN: %clang_cc1 %s -pedantic -ast-dump -verify -triple i386-windows | FileCheck %s
//CHECK: CXXMethodDecl {{.*}} constexpr operator() 'int (__private int){{.*}} const __generic'
auto glambda = [](auto a) { return a; };
-//RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify
+//RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify
struct C {
auto fGlob() __global -> decltype(this);
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify -fsyntax-only
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify -fsyntax-only
// expected-no-diagnostics
// Extract from PR38614
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify -fsyntax-only
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify -fsyntax-only
__global const int& f(__global float &ref) {
return ref; // expected-error{{reference of type 'const __global int &' cannot bind to a temporary object because of address space mismatch}}
-//RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -verify -ast-dump | FileCheck %s
+//RUN: %clang_cc1 %s -pedantic -verify -ast-dump | FileCheck %s
template <typename T>
struct S {
-// RUN: %clang_cc1 %s -verify -pedantic -fsyntax-only -cl-std=clc++
+// RUN: %clang_cc1 %s -verify -pedantic -fsyntax-only
// expected-no-diagnostics
struct RetGlob {
-//RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -ast-dump -verify | FileCheck %s
+//RUN: %clang_cc1 %s -pedantic -ast-dump -verify | FileCheck %s
__constant int i = 1;
//CHECK: |-VarDecl {{.*}} ai '__global int':'__global int'
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify -fsyntax-only
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify -fsyntax-only
void foo(global int *gl, const global int *gl_const, global int &gl_ref) {
//FIXME: Diagnostics can be improved to be more specific in some cases.
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify -ast-dump | FileCheck %s
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify -ast-dump | FileCheck %s
// expected-no-diagnostics
-// RUN: %clang_cc1 %s -cl-std=clc++ -pedantic -verify -fsyntax-only
+// RUN: %clang_cc1 %s -pedantic -verify -fsyntax-only
struct C {
kernel void m(); //expected-error{{kernel functions cannot be class members}}
-//RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -verify -fsyntax-only
-//RUN: %clang_cc1 %s -triple spir -cl-std=clc++ -verify -fsyntax-only -DFUNCPTREXT
+//RUN: %clang_cc1 %s -triple spir -verify -fsyntax-only
+//RUN: %clang_cc1 %s -triple spir -verify -fsyntax-only -DFUNCPTREXT
#ifdef FUNCPTREXT
#pragma OPENCL EXTENSION __cl_clang_function_pointers : enable
-//RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify
+//RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify
struct C {
void m1() __local __local; //expected-warning{{multiple identical address spaces specified for type}}
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify -fsyntax-only
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify -fsyntax-only
class A {
public:
-//RUN: %clang_cc1 %s -cl-std=clc++ -verify -fsyntax-only -triple spir
-//RUN: %clang_cc1 %s -cl-std=clc++ -verify -fsyntax-only -DFPTREXT -triple spir
+//RUN: %clang_cc1 %s -verify -fsyntax-only -triple spir
+//RUN: %clang_cc1 %s -verify -fsyntax-only -DFPTREXT -triple spir
#ifdef FPTREXT
#pragma OPENCL EXTENSION __cl_clang_function_pointers : enable
-// RUN: %clang_cc1 %s -triple spir-unknown-unknown -cl-std=clc++ -pedantic -verify -fsyntax-only
+// RUN: %clang_cc1 %s -triple spir-unknown-unknown -pedantic -verify -fsyntax-only
// This test checks that various C/C++/OpenCL C constructs are not available in
// C++ for OpenCL.
# suffixes: A list of file extensions to treat as test files.
config.suffixes = ['.c', '.cpp', '.i', '.cppm', '.m', '.mm', '.cu',
- '.ll', '.cl', '.s', '.S', '.modulemap', '.test', '.rs', '.ifs']
+ '.ll', '.cl', '.clcpp', '.s', '.S', '.modulemap', '.test', '.rs', '.ifs']
# excludes: A list of directories to exclude from the testsuite. The 'Inputs'
# subdirectories contain auxiliary inputs for various tests in their parent