microsoft
/
clang-1
зеркало из https://github.com/microsoft/clang-1.git
1
0
Форкнуть 0
Код Задачи Пакеты Проекты Релизы Вики Активность

Re-commit r170428 changes with Linux style file endings. 

Add OpenCL images as clang builtin types.

git-svn-id: https://llvm.org/svn/llvm-project/cfe/trunk@170432 91177308-0d34-0410-b5e6-96231b3b80d8
This commit is contained in:
Guy Benyei 2012-12-18 14:38:23 +00:00
Родитель 7f92f2d8d9
Коммит b13621d08e
33 изменённых файлов: 416 добавлений и 7 удалений
Показать статистику Скачать Patch файл Скачать Diff файл Показать все файлы Свернуть все файлы

3
include/clang/AST/ASTContext.h
Просмотреть файл

@ -716,6 +716,9 @@ public:
CanQualType PseudoObjectTy, ARCUnbridgedCastTy;
CanQualType ObjCBuiltinIdTy, ObjCBuiltinClassTy, ObjCBuiltinSelTy;
CanQualType ObjCBuiltinBoolTy;
CanQualType OCLImage1dTy, OCLImage1dArrayTy, OCLImage1dBufferTy;
CanQualType OCLImage2dTy, OCLImage2dArrayTy;
CanQualType OCLImage3dTy;
// Types for deductions in C++0x [stmt.ranged]'s desugaring. Built on demand.
mutable QualType AutoDeductTy; // Deduction against 'auto'.

8
include/clang/AST/BuiltinTypes.def
Просмотреть файл

@ -154,6 +154,14 @@ BUILTIN_TYPE(ObjCClass, ObjCBuiltinClassTy)
// type is a typedef of a PointerType to this.
BUILTIN_TYPE(ObjCSel, ObjCBuiltinSelTy)
// OpenCL image types.
BUILTIN_TYPE(OCLImage1d, OCLImage1dTy)
BUILTIN_TYPE(OCLImage1dArray, OCLImage1dArrayTy)
BUILTIN_TYPE(OCLImage1dBuffer, OCLImage1dBufferTy)
BUILTIN_TYPE(OCLImage2d, OCLImage2dTy)
BUILTIN_TYPE(OCLImage2dArray, OCLImage2dArrayTy)
BUILTIN_TYPE(OCLImage3d, OCLImage3dTy)
// This represents the type of an expression whose type is
// totally unknown, e.g. 'T::foo'. It is permitted for this to
// appear in situations where the structure of the type is

45
include/clang/AST/Type.h
Просмотреть файл

@ -1580,6 +1580,17 @@ public:
bool isNullPtrType() const; // C++0x nullptr_t
bool isAtomicType() const; // C11 _Atomic()
bool isImage1dT() const; // OpenCL image1d_t
bool isImage1dArrayT() const; // OpenCL image1d_array_t
bool isImage1dBufferT() const; // OpenCL image1d_buffer_t
bool isImage2dT() const; // OpenCL image2d_t
bool isImage2dArrayT() const; // OpenCL image2d_array_t
bool isImage3dT() const; // OpenCL image3d_t
bool isImageType() const; // Any OpenCL image type
bool isOpenCLSpecificType() const; // Any OpenCL specific type
/// Determines if this type, which must satisfy
/// isObjCLifetimeType(), is implicitly __unsafe_unretained rather
/// than implicitly __strong.
@ -4883,6 +4894,40 @@ inline bool Type::isObjCSelType() const {
inline bool Type::isObjCBuiltinType() const {
return isObjCIdType() || isObjCClassType() || isObjCSelType();
}
inline bool Type::isImage1dT() const {
return isSpecificBuiltinType(BuiltinType::OCLImage1d);
}
inline bool Type::isImage1dArrayT() const {
return isSpecificBuiltinType(BuiltinType::OCLImage1dArray);
}
inline bool Type::isImage1dBufferT() const {
return isSpecificBuiltinType(BuiltinType::OCLImage1dBuffer);
}
inline bool Type::isImage2dT() const {
return isSpecificBuiltinType(BuiltinType::OCLImage2d);
}
inline bool Type::isImage2dArrayT() const {
return isSpecificBuiltinType(BuiltinType::OCLImage2dArray);
}
inline bool Type::isImage3dT() const {
return isSpecificBuiltinType(BuiltinType::OCLImage3d);
}
inline bool Type::isImageType() const {
return isImage3dT() ||
isImage2dT() || isImage2dArrayT() ||
isImage1dT() || isImage1dArrayT() || isImage1dBufferT();
}
inline bool Type::isOpenCLSpecificType() const {
return isImageType();
}
inline bool Type::isTemplateTypeParmType() const {
return isa<TemplateTypeParmType>(CanonicalType);
}

8
include/clang/Basic/Specifiers.h
Просмотреть файл

@ -62,13 +62,19 @@ namespace clang {
TST_auto, // C++0x auto
TST_unknown_anytype, // __unknown_anytype extension
TST_atomic, // C11 _Atomic
TST_image1d_t, // OpenCL image1d_t
TST_image1d_array_t, // OpenCL image1d_array_t
TST_image1d_buffer_t, // OpenCL image1d_buffer_t
TST_image2d_t, // OpenCL image2d_t
TST_image2d_array_t, // OpenCL image2d_array_t
TST_image3d_t, // OpenCL image3d_t
TST_error // erroneous type
};
/// \brief Structure that packs information about the type specifiers that
/// were written in a particular type specifier sequence.
struct WrittenBuiltinSpecs {
/*DeclSpec::TST*/ unsigned Type : 5;
/*DeclSpec::TST*/ unsigned Type : 6;
/*DeclSpec::TSS*/ unsigned Sign : 2;
/*DeclSpec::TSW*/ unsigned Width : 2;
bool ModeAttr : 1;

6
include/clang/Basic/TokenKinds.def
Просмотреть файл

@ -448,6 +448,12 @@ ALIAS("read_only", __read_only , KEYOPENCL)
ALIAS("write_only", __write_only , KEYOPENCL)
ALIAS("read_write", __read_write , KEYOPENCL)
KEYWORD(__builtin_astype , KEYOPENCL)
KEYWORD(image1d_t , KEYOPENCL)
KEYWORD(image1d_array_t , KEYOPENCL)
KEYWORD(image1d_buffer_t , KEYOPENCL)
KEYWORD(image2d_t , KEYOPENCL)
KEYWORD(image2d_array_t , KEYOPENCL)
KEYWORD(image3d_t , KEYOPENCL)
// Borland Extensions.
KEYWORD(__pascal , KEYALL)

8
include/clang/Sema/DeclSpec.h
Просмотреть файл

@ -276,6 +276,12 @@ public:
static const TST TST_auto = clang::TST_auto;
static const TST TST_unknown_anytype = clang::TST_unknown_anytype;
static const TST TST_atomic = clang::TST_atomic;
static const TST TST_image1d_t = clang::TST_image1d_t;
static const TST TST_image1d_array_t = clang::TST_image1d_array_t;
static const TST TST_image1d_buffer_t = clang::TST_image1d_buffer_t;
static const TST TST_image2d_t = clang::TST_image2d_t;
static const TST TST_image2d_array_t = clang::TST_image2d_array_t;
static const TST TST_image3d_t = clang::TST_image3d_t;
static const TST TST_error = clang::TST_error;
// type-qualifiers
@ -306,7 +312,7 @@ private:
/*TSW*/unsigned TypeSpecWidth : 2;
/*TSC*/unsigned TypeSpecComplex : 2;
/*TSS*/unsigned TypeSpecSign : 2;
/*TST*/unsigned TypeSpecType : 5;
/*TST*/unsigned TypeSpecType : 6;
unsigned TypeAltiVecVector : 1;
unsigned TypeAltiVecPixel : 1;
unsigned TypeAltiVecBool : 1;

14
include/clang/Serialization/ASTBitCodes.h
Просмотреть файл

@ -701,7 +701,19 @@ namespace clang {
/// \brief The __va_list_tag placeholder type.
PREDEF_TYPE_VA_LIST_TAG = 36,
/// \brief The placeholder type for builtin functions.
PREDEF_TYPE_BUILTIN_FN = 37
PREDEF_TYPE_BUILTIN_FN = 37,
/// \brief OpenCL 1d image type.
PREDEF_TYPE_IMAGE1D_ID = 38,
/// \brief OpenCL 1d image array type.
PREDEF_TYPE_IMAGE1D_ARR_ID = 39,
/// \brief OpenCL 1d image buffer type.
PREDEF_TYPE_IMAGE1D_BUFF_ID = 40,
/// \brief OpenCL 2d image type.
PREDEF_TYPE_IMAGE2D_ID = 41,
/// \brief OpenCL 2d image array type.
PREDEF_TYPE_IMAGE2D_ARR_ID = 42,
/// \brief OpenCL 3d image type.
PREDEF_TYPE_IMAGE3D_ID = 43
};
/// \brief The number of predefined type IDs that are reserved for

19
lib/AST/ASTContext.cpp
Просмотреть файл

@ -875,6 +875,15 @@ void ASTContext::InitBuiltinTypes(const TargetInfo &Target) {
InitBuiltinType(ObjCBuiltinClassTy, BuiltinType::ObjCClass);
InitBuiltinType(ObjCBuiltinSelTy, BuiltinType::ObjCSel);
if (LangOpts.OpenCL) {
InitBuiltinType(OCLImage1dTy, BuiltinType::OCLImage1d);
InitBuiltinType(OCLImage1dArrayTy, BuiltinType::OCLImage1dArray);
InitBuiltinType(OCLImage1dBufferTy, BuiltinType::OCLImage1dBuffer);
InitBuiltinType(OCLImage2dTy, BuiltinType::OCLImage2d);
InitBuiltinType(OCLImage2dArrayTy, BuiltinType::OCLImage2dArray);
InitBuiltinType(OCLImage3dTy, BuiltinType::OCLImage3d);
}
// Builtin type for __objc_yes and __objc_no
ObjCBuiltinBoolTy = (Target.useSignedCharForObjCBool() ?
SignedCharTy : BoolTy);
@ -1412,6 +1421,16 @@ ASTContext::getTypeInfoImpl(const Type *T) const {
Width = Target->getPointerWidth(0);
Align = Target->getPointerAlign(0);
break;
case BuiltinType::OCLImage1d:
case BuiltinType::OCLImage1dArray:
case BuiltinType::OCLImage1dBuffer:
case BuiltinType::OCLImage2d:
case BuiltinType::OCLImage2dArray:
case BuiltinType::OCLImage3d:
// Currently these types are pointers to opaque types.
Width = Target->getPointerWidth(0);
Align = Target->getPointerAlign(0);
break;
}
break;
case Type::ObjCObjectPointer:

6
lib/AST/ItaniumMangle.cpp
Просмотреть файл

@ -1880,6 +1880,12 @@ void CXXNameMangler::mangleType(const BuiltinType *T) {
case BuiltinType::ObjCId: Out << "11objc_object"; break;
case BuiltinType::ObjCClass: Out << "10objc_class"; break;
case BuiltinType::ObjCSel: Out << "13objc_selector"; break;
case BuiltinType::OCLImage1d: Out << "11ocl_image1d"; break;
case BuiltinType::OCLImage1dArray: Out << "16ocl_image1darray"; break;
case BuiltinType::OCLImage1dBuffer: Out << "17ocl_image1dbuffer"; break;
case BuiltinType::OCLImage2d: Out << "11ocl_image2d"; break;
case BuiltinType::OCLImage2dArray: Out << "16ocl_image2darray"; break;
case BuiltinType::OCLImage3d: Out << "11ocl_image3d"; break;
}
}

7
lib/AST/MicrosoftMangle.cpp
Просмотреть файл

@ -1054,6 +1054,13 @@ void MicrosoftCXXNameMangler::mangleType(const BuiltinType *T,
case BuiltinType::ObjCClass: Out << "PAUobjc_class@@"; break;
case BuiltinType::ObjCSel: Out << "PAUobjc_selector@@"; break;
case BuiltinType::OCLImage1d: Out << "PAUocl_image1d@@"; break;
case BuiltinType::OCLImage1dArray: Out << "PAUocl_image1darray@@"; break;
case BuiltinType::OCLImage1dBuffer: Out << "PAUocl_image1dbuffer@@"; break;
case BuiltinType::OCLImage2d: Out << "PAUocl_image2d@@"; break;
case BuiltinType::OCLImage2dArray: Out << "PAUocl_image2darray@@"; break;
case BuiltinType::OCLImage3d: Out << "PAUocl_image3d@@"; break;
case BuiltinType::NullPtr: Out << "$$T"; break;
case BuiltinType::Char16:

6
lib/AST/NSAPI.cpp
Просмотреть файл

@ -337,6 +337,12 @@ NSAPI::getNSNumberFactoryMethodKind(QualType T) const {
case BuiltinType::ObjCClass:
case BuiltinType::ObjCId:
case BuiltinType::ObjCSel:
case BuiltinType::OCLImage1d:
case BuiltinType::OCLImage1dArray:
case BuiltinType::OCLImage1dBuffer:
case BuiltinType::OCLImage2d:
case BuiltinType::OCLImage2dArray:
case BuiltinType::OCLImage3d:
case BuiltinType::BoundMember:
case BuiltinType::Dependent:
case BuiltinType::Overload:

6
lib/AST/Type.cpp
Просмотреть файл

@ -1512,6 +1512,12 @@ StringRef BuiltinType::getName(const PrintingPolicy &Policy) const {
case ObjCId: return "id";
case ObjCClass: return "Class";
case ObjCSel: return "SEL";
case OCLImage1d: return "image1d_t";
case OCLImage1dArray: return "image1d_array_t";
case OCLImage1dBuffer: return "image1d_buffer_t";
case OCLImage2d: return "image2d_t";
case OCLImage2dArray: return "image2d_array_t";
case OCLImage3d: return "image3d_t";
}
llvm_unreachable("Invalid builtin type.");

6
lib/AST/TypeLoc.cpp
Просмотреть файл

@ -262,6 +262,12 @@ TypeSpecifierType BuiltinTypeLoc::getWrittenTypeSpec() const {
case BuiltinType::ObjCId:
case BuiltinType::ObjCClass:
case BuiltinType::ObjCSel:
case BuiltinType::OCLImage1d:
case BuiltinType::OCLImage1dArray:
case BuiltinType::OCLImage1dBuffer:
case BuiltinType::OCLImage2d:
case BuiltinType::OCLImage2dArray:
case BuiltinType::OCLImage3d:
case BuiltinType::BuiltinFn:
return TST_unspecified;
}

32
lib/CodeGen/CGDebugInfo.cpp
Просмотреть файл

@ -407,6 +407,26 @@ llvm::DIType CGDebugInfo::CreateType(const BuiltinType *BT) {
0);
return SelTy;
}
case BuiltinType::OCLImage1d:
return getOrCreateStructPtrType("opencl_image1d_t",
OCLImage1dDITy);
case BuiltinType::OCLImage1dArray:
return getOrCreateStructPtrType("opencl_image1d_array_t",
OCLImage1dArrayDITy);
case BuiltinType::OCLImage1dBuffer:
return getOrCreateStructPtrType("opencl_image1d_buffer_t",
OCLImage1dBufferDITy);
case BuiltinType::OCLImage2d:
return getOrCreateStructPtrType("opencl_image2d_t",
OCLImage2dDITy);
case BuiltinType::OCLImage2dArray:
return getOrCreateStructPtrType("opencl_image2d_array_t",
OCLImage2dArrayDITy);
case BuiltinType::OCLImage3d:
return getOrCreateStructPtrType("opencl_image3d_t",
OCLImage3dDITy);
case BuiltinType::UChar:
case BuiltinType::Char_U: Encoding = llvm::dwarf::DW_ATE_unsigned_char; break;
case BuiltinType::Char_S:
@ -613,6 +633,18 @@ llvm::DIType CGDebugInfo::CreatePointerLikeType(unsigned Tag,
Size, Align);
}
llvm::DIType CGDebugInfo::getOrCreateStructPtrType(StringRef Name, llvm::DIType &Cache) {
if (Cache.Verify())
return Cache;
Cache =
DBuilder.createForwardDecl(llvm::dwarf::DW_TAG_structure_type,
Name, TheCU, getOrCreateMainFile(),
0);
unsigned Size = CGM.getContext().getTypeSize(CGM.getContext().VoidPtrTy);
Cache = DBuilder.createPointerType(Cache, Size);
return Cache;
}
llvm::DIType CGDebugInfo::CreateType(const BlockPointerType *Ty,
llvm::DIFile Unit) {
if (BlockLiteralGenericSet)

5
lib/CodeGen/CGDebugInfo.h
Просмотреть файл

@ -52,6 +52,9 @@ class CGDebugInfo {
llvm::DIType ClassTy;
llvm::DIType ObjTy;
llvm::DIType SelTy;
llvm::DIType OCLImage1dDITy, OCLImage1dArrayDITy, OCLImage1dBufferDITy;
llvm::DIType OCLImage2dDITy, OCLImage2dArrayDITy;
llvm::DIType OCLImage3dDITy;
/// TypeCache - Cache of previously constructed Types.
llvm::DenseMap<void *, llvm::WeakVH> TypeCache;
@ -117,6 +120,8 @@ class CGDebugInfo {
const Type *Ty, QualType PointeeTy,
llvm::DIFile F);
llvm::DIType getOrCreateStructPtrType(StringRef Name, llvm::DIType &Cache);
llvm::DISubprogram CreateCXXMemberFunction(const CXXMethodDecl *Method,
llvm::DIFile F,
llvm::DIType RecordTy);

31
lib/CodeGen/CGOpenCLRuntime.cpp
Просмотреть файл

@ -16,6 +16,8 @@
#include "CGOpenCLRuntime.h"
#include "CodeGenFunction.h"
#include "llvm/GlobalValue.h"
#include "llvm/DerivedTypes.h"
#include <assert.h>
using namespace clang;
using namespace CodeGen;
@ -26,3 +28,32 @@ void CGOpenCLRuntime::EmitWorkGroupLocalVarDecl(CodeGenFunction &CGF,
const VarDecl &D) {
return CGF.EmitStaticVarDecl(D, llvm::GlobalValue::InternalLinkage);
}
llvm::Type *CGOpenCLRuntime::convertOpenCLSpecificType(const Type *T) {
assert(T->isOpenCLSpecificType() &&
"Not an OpenCL specific type!");
switch (cast<BuiltinType>(T)->getKind()) {
default:
llvm_unreachable("Unexpected opencl builtin type!");
return 0;
case BuiltinType::OCLImage1d:
return llvm::PointerType::get(llvm::StructType::create(
CGM.getLLVMContext(), "opencl.image1d_t"), 0);
case BuiltinType::OCLImage1dArray:
return llvm::PointerType::get(llvm::StructType::create(
CGM.getLLVMContext(), "opencl.image1d_array_t"), 0);
case BuiltinType::OCLImage1dBuffer:
return llvm::PointerType::get(llvm::StructType::create(
CGM.getLLVMContext(), "opencl.image1d_buffer_t"), 0);
case BuiltinType::OCLImage2d:
return llvm::PointerType::get(llvm::StructType::create(
CGM.getLLVMContext(), "opencl.image2d_t"), 0);
case BuiltinType::OCLImage2dArray:
return llvm::PointerType::get(llvm::StructType::create(
CGM.getLLVMContext(), "opencl.image2d_array_t"), 0);
case BuiltinType::OCLImage3d:
return llvm::PointerType::get(llvm::StructType::create(
CGM.getLLVMContext(), "opencl.image3d_t"), 0);
}
}

6
lib/CodeGen/CGOpenCLRuntime.h
Просмотреть файл

@ -16,6 +16,10 @@
#ifndef CLANG_CODEGEN_OPENCLRUNTIME_H
#define CLANG_CODEGEN_OPENCLRUNTIME_H
#include "clang/AST/Type.h"
#include "llvm/Type.h"
#include "llvm/Value.h"
namespace clang {
class VarDecl;
@ -38,6 +42,8 @@ public:
/// CodeGenFunction::EmitStaticVarDecl to emit an internal global for D.
virtual void EmitWorkGroupLocalVarDecl(CodeGenFunction &CGF,
const VarDecl &D);
virtual llvm::Type *convertOpenCLSpecificType(const Type *T);
};
}

6
lib/CodeGen/CGRTTI.cpp
Просмотреть файл

@ -191,6 +191,12 @@ static bool TypeInfoIsInStandardLibrary(const BuiltinType *Ty) {
case BuiltinType::Char32:
case BuiltinType::Int128:
case BuiltinType::UInt128:
case BuiltinType::OCLImage1d:
case BuiltinType::OCLImage1dArray:
case BuiltinType::OCLImage1dBuffer:
case BuiltinType::OCLImage2d:
case BuiltinType::OCLImage2dArray:
case BuiltinType::OCLImage3d:
return true;
case BuiltinType::Dependent:

10
lib/CodeGen/CodeGenTypes.cpp
Просмотреть файл

@ -14,6 +14,7 @@
#include "CodeGenTypes.h"
#include "CGCXXABI.h"
#include "CGCall.h"
#include "CGOpenCLRuntime.h"
#include "CGRecordLayout.h"
#include "TargetInfo.h"
#include "clang/AST/ASTContext.h"
@ -367,6 +368,15 @@ llvm::Type *CodeGenTypes::ConvertType(QualType T) {
ResultType = llvm::IntegerType::get(getLLVMContext(), 128);
break;
case BuiltinType::OCLImage1d:
case BuiltinType::OCLImage1dArray:
case BuiltinType::OCLImage1dBuffer:
case BuiltinType::OCLImage2d:
case BuiltinType::OCLImage2dArray:
case BuiltinType::OCLImage3d:
ResultType = CGM.getOpenCLRuntime().convertOpenCLSpecificType(Ty);
break;
case BuiltinType::Dependent:
#define BUILTIN_TYPE(Id, SingletonId)
#define PLACEHOLDER_TYPE(Id, SingletonId) \

48
lib/Parse/ParseDecl.cpp
Просмотреть файл

@ -2753,6 +2753,30 @@ void Parser::ParseDeclarationSpecifiers(DeclSpec &DS,
case tok::kw___pixel:
isInvalid = DS.SetTypeAltiVecPixel(true, Loc, PrevSpec, DiagID);
break;
case tok::kw_image1d_t:
isInvalid = DS.SetTypeSpecType(DeclSpec::TST_image1d_t, Loc,
PrevSpec, DiagID);
break;
case tok::kw_image1d_array_t:
isInvalid = DS.SetTypeSpecType(DeclSpec::TST_image1d_array_t, Loc,
PrevSpec, DiagID);
break;
case tok::kw_image1d_buffer_t:
isInvalid = DS.SetTypeSpecType(DeclSpec::TST_image1d_buffer_t, Loc,
PrevSpec, DiagID);
break;
case tok::kw_image2d_t:
isInvalid = DS.SetTypeSpecType(DeclSpec::TST_image2d_t, Loc,
PrevSpec, DiagID);
break;
case tok::kw_image2d_array_t:
isInvalid = DS.SetTypeSpecType(DeclSpec::TST_image2d_array_t, Loc,
PrevSpec, DiagID);
break;
case tok::kw_image3d_t:
isInvalid = DS.SetTypeSpecType(DeclSpec::TST_image3d_t, Loc,
PrevSpec, DiagID);
break;
case tok::kw___unknown_anytype:
isInvalid = DS.SetTypeSpecType(TST_unknown_anytype, Loc,
PrevSpec, DiagID);
@ -3596,6 +3620,14 @@ bool Parser::isKnownToBeTypeSpecifier(const Token &Tok) const {
case tok::kw__Decimal128:
case tok::kw___vector:
// OpenCL specific types:
case tok::kw_image1d_t:
case tok::kw_image1d_array_t:
case tok::kw_image1d_buffer_t:
case tok::kw_image2d_t:
case tok::kw_image2d_array_t:
case tok::kw_image3d_t:
// struct-or-union-specifier (C99) or class-specifier (C++)
case tok::kw_class:
case tok::kw_struct:
@ -3668,6 +3700,14 @@ bool Parser::isTypeSpecifierQualifier() {
case tok::kw__Decimal128:
case tok::kw___vector:
// OpenCL specific types:
case tok::kw_image1d_t:
case tok::kw_image1d_array_t:
case tok::kw_image1d_buffer_t:
case tok::kw_image2d_t:
case tok::kw_image2d_array_t:
case tok::kw_image3d_t:
// struct-or-union-specifier (C99) or class-specifier (C++)
case tok::kw_class:
case tok::kw_struct:
@ -3812,6 +3852,14 @@ bool Parser::isDeclarationSpecifier(bool DisambiguatingWithExpression) {
case tok::kw__Decimal128:
case tok::kw___vector:
// OpenCL specific types:
case tok::kw_image1d_t:
case tok::kw_image1d_array_t:
case tok::kw_image1d_buffer_t:
case tok::kw_image2d_t:
case tok::kw_image2d_array_t:
case tok::kw_image3d_t:
// struct-or-union-specifier (C99) or class-specifier (C++)
case tok::kw_class:
case tok::kw_struct:

8
lib/Parse/ParseExpr.cpp
Просмотреть файл

@ -1078,7 +1078,13 @@ ExprResult Parser::ParseCastExpression(bool isUnaryExpression,
case tok::kw_void:
case tok::kw_typename:
case tok::kw_typeof:
case tok::kw___vector: {
case tok::kw___vector:
case tok::kw_image1d_t:
case tok::kw_image1d_array_t:
case tok::kw_image1d_buffer_t:
case tok::kw_image2d_t:
case tok::kw_image2d_array_t:
case tok::kw_image3d_t: {
if (!getLangOpts().CPlusPlus) {
Diag(Tok, diag::err_expected_expression);
return ExprError();

6
lib/Parse/ParseTentative.cpp
Просмотреть файл

@ -837,6 +837,12 @@ Parser::isExpressionOrTypeSpecifierSimple(tok::TokenKind Kind) {
case tok::kw___vector:
case tok::kw___pixel:
case tok::kw__Atomic:
case tok::kw_image1d_t:
case tok::kw_image1d_array_t:
case tok::kw_image1d_buffer_t:
case tok::kw_image2d_t:
case tok::kw_image2d_array_t:
case tok::kw_image3d_t:
case tok::kw___unknown_anytype:
return TPResult::False();

12
lib/Sema/DeclSpec.cpp
Просмотреть файл

@ -280,6 +280,12 @@ bool Declarator::isDeclarationOfFunction() const {
case TST_unspecified:
case TST_void:
case TST_wchar:
case TST_image1d_t:
case TST_image1d_array_t:
case TST_image1d_buffer_t:
case TST_image2d_t:
case TST_image2d_array_t:
case TST_image3d_t:
return false;
case TST_decltype:
@ -414,6 +420,12 @@ const char *DeclSpec::getSpecifierName(DeclSpec::TST T) {
case DeclSpec::TST_underlyingType: return "__underlying_type";
case DeclSpec::TST_unknown_anytype: return "__unknown_anytype";
case DeclSpec::TST_atomic: return "_Atomic";
case DeclSpec::TST_image1d_t: return "image1d_t";
case DeclSpec::TST_image1d_array_t: return "image1d_array_t";
case DeclSpec::TST_image1d_buffer_t: return "image1d_buffer_t";
case DeclSpec::TST_image2d_t: return "image2d_t";
case DeclSpec::TST_image2d_array_t: return "image2d_array_t";
case DeclSpec::TST_image3d_t: return "image3d_t";
case DeclSpec::TST_error: return "(error)";
}
llvm_unreachable("Unknown typespec!");

6
lib/Sema/SemaTemplateVariadic.cpp
Просмотреть файл

@ -731,6 +731,12 @@ bool Sema::containsUnexpandedParameterPacks(Declarator &D) {
case TST_class:
case TST_auto:
case TST_unknown_anytype:
case TST_image1d_t:
case TST_image1d_array_t:
case TST_image1d_buffer_t:
case TST_image2d_t:
case TST_image2d_array_t:
case TST_image3d_t:
case TST_error:
break;
}

24
lib/Sema/SemaType.cpp
Просмотреть файл

@ -903,6 +903,30 @@ static QualType ConvertDeclSpecToType(TypeProcessingState &state) {
}
break;
case DeclSpec::TST_image1d_t:
Result = Context.OCLImage1dTy;
break;
case DeclSpec::TST_image1d_array_t:
Result = Context.OCLImage1dArrayTy;
break;
case DeclSpec::TST_image1d_buffer_t:
Result = Context.OCLImage1dBufferTy;
break;
case DeclSpec::TST_image2d_t:
Result = Context.OCLImage2dTy;
break;
case DeclSpec::TST_image2d_array_t:
Result = Context.OCLImage2dArrayTy;
break;
case DeclSpec::TST_image3d_t:
Result = Context.OCLImage3dTy;
break;
case DeclSpec::TST_error:
Result = Context.IntTy;
declarator.setInvalidType(true);

6
lib/Serialization/ASTCommon.cpp
Просмотреть файл

@ -60,6 +60,12 @@ serialization::TypeIdxFromBuiltin(const BuiltinType *BT) {
case BuiltinType::ObjCId: ID = PREDEF_TYPE_OBJC_ID; break;
case BuiltinType::ObjCClass: ID = PREDEF_TYPE_OBJC_CLASS; break;
case BuiltinType::ObjCSel: ID = PREDEF_TYPE_OBJC_SEL; break;
case BuiltinType::OCLImage1d: ID = PREDEF_TYPE_IMAGE1D_ID; break;
case BuiltinType::OCLImage1dArray: ID = PREDEF_TYPE_IMAGE1D_ARR_ID; break;
case BuiltinType::OCLImage1dBuffer: ID = PREDEF_TYPE_IMAGE1D_BUFF_ID; break;
case BuiltinType::OCLImage2d: ID = PREDEF_TYPE_IMAGE2D_ID; break;
case BuiltinType::OCLImage2dArray: ID = PREDEF_TYPE_IMAGE2D_ARR_ID; break;
case BuiltinType::OCLImage3d: ID = PREDEF_TYPE_IMAGE3D_ID; break;
case BuiltinType::BuiltinFn:
ID = PREDEF_TYPE_BUILTIN_FN; break;

6
lib/Serialization/ASTReader.cpp
Просмотреть файл

@ -4883,6 +4883,12 @@ QualType ASTReader::GetType(TypeID ID) {
case PREDEF_TYPE_OBJC_ID: T = Context.ObjCBuiltinIdTy; break;
case PREDEF_TYPE_OBJC_CLASS: T = Context.ObjCBuiltinClassTy; break;
case PREDEF_TYPE_OBJC_SEL: T = Context.ObjCBuiltinSelTy; break;
case PREDEF_TYPE_IMAGE1D_ID: T = Context.OCLImage1dTy; break;
case PREDEF_TYPE_IMAGE1D_ARR_ID: T = Context.OCLImage1dArrayTy; break;
case PREDEF_TYPE_IMAGE1D_BUFF_ID: T = Context.OCLImage1dBufferTy; break;
case PREDEF_TYPE_IMAGE2D_ID: T = Context.OCLImage2dTy; break;
case PREDEF_TYPE_IMAGE2D_ARR_ID: T = Context.OCLImage2dArrayTy; break;
case PREDEF_TYPE_IMAGE3D_ID: T = Context.OCLImage3dTy; break;
case PREDEF_TYPE_AUTO_DEDUCT: T = Context.getAutoDeductType(); break;
case PREDEF_TYPE_AUTO_RREF_DEDUCT:

22
test/CodeGenOpenCL/opencl_types.cl Normal file
Просмотреть файл

@ -0,0 +1,22 @@
// RUN: %clang_cc1 %s -emit-llvm -o - -O0 | FileCheck %s
void fnc1(image1d_t img) {}
// CHECK: @fnc1(%opencl.image1d_t*
void fnc1arr(image1d_array_t img) {}
// CHECK: @fnc1arr(%opencl.image1d_array_t*
void fnc1buff(image1d_buffer_t img) {}
// CHECK: @fnc1buff(%opencl.image1d_buffer_t*
void fnc2(image2d_t img) {}
// CHECK: @fnc2(%opencl.image2d_t*
void fnc2arr(image2d_array_t img) {}
// CHECK: @fnc2arr(%opencl.image2d_array_t*
void fnc3(image3d_t img) {}
// CHECK: @fnc3(%opencl.image3d_t*
kernel void foo(image1d_t img) {
}

18
test/PCH/ocl_types.cl Normal file
Просмотреть файл

@ -0,0 +1,18 @@
// Test this without pch.
// RUN: %clang_cc1 -include %S/ocl_types.h -fsyntax-only %s
// Test with pch.
// RUN: %clang_cc1 -x cl -emit-pch -o %t %S/ocl_types.h
// RUN: %clang_cc1 -include-pch %t -fsyntax-only %s -ast-print
void foo1(img1d_t img);
void foo2(img1darr_t img);
void foo3(img1dbuff_t img);
void foo4(img2d_t img);
void foo5(img2darr_t img);
void foo6(img3d_t img);

19
test/PCH/ocl_types.h Normal file
Просмотреть файл

@ -0,0 +1,19 @@
/* Used with the ocl_types.cl test */
// image1d_t
typedef image1d_t img1d_t;
// image1d_array_t
typedef image1d_array_t img1darr_t;
// image1d_buffer_t
typedef image1d_buffer_t img1dbuff_t;
// image2d_t
typedef image2d_t img2d_t;
// image2d_array_t
typedef image2d_array_t img2darr_t;
// image3d_t
typedef image3d_t img3d_t;

2
test/Parser/opencl-image-access.cl
Просмотреть файл

@ -1,7 +1,5 @@
// RUN: %clang_cc1 %s -fsyntax-only
typedef void* image2d_t;
__kernel void f__ro(__read_only image2d_t a) { }
__kernel void f__wo(__write_only image2d_t a) { }

6
tools/libclang/CIndex.cpp
Просмотреть файл

@ -1367,6 +1367,12 @@ bool CursorVisitor::VisitBuiltinTypeLoc(BuiltinTypeLoc TL) {
case BuiltinType::Void:
case BuiltinType::NullPtr:
case BuiltinType::Dependent:
case BuiltinType::OCLImage1d:
case BuiltinType::OCLImage1dArray:
case BuiltinType::OCLImage1dBuffer:
case BuiltinType::OCLImage2d:
case BuiltinType::OCLImage2dArray:
case BuiltinType::OCLImage3d:
#define BUILTIN_TYPE(Id, SingletonId)
#define SIGNED_TYPE(Id, SingletonId) case BuiltinType::Id:
#define UNSIGNED_TYPE(Id, SingletonId) case BuiltinType::Id:

6
tools/libclang/CIndexUSRs.cpp
Просмотреть файл

@ -587,6 +587,12 @@ void USRGenerator::VisitType(QualType T) {
#define PLACEHOLDER_TYPE(Id, SingletonId) case BuiltinType::Id:
#include "clang/AST/BuiltinTypes.def"
case BuiltinType::Dependent:
case BuiltinType::OCLImage1d:
case BuiltinType::OCLImage1dArray:
case BuiltinType::OCLImage1dBuffer:
case BuiltinType::OCLImage2d:
case BuiltinType::OCLImage2dArray:
case BuiltinType::OCLImage3d:
IgnoreResults = true;
return;
case BuiltinType::ObjCId: