From 2b2e574cc3c0bdcc18b4cc56a38dc63d7ff32efd Mon Sep 17 00:00:00 2001 From: Artem Belevich Date: Thu, 19 Mar 2015 17:32:06 +0000 Subject: [PATCH] Ensure that we still parse preprocessed CUDA files as CUDA when we use -save-temps option. Summary: Fixes PR22926. Review: http://reviews.llvm.org/D8383 git-svn-id: https://llvm.org/svn/llvm-project/cfe/trunk@232737 91177308-0d34-0410-b5e6-96231b3b80d8 --- include/clang/Driver/Types.def | 3 ++- include/clang/Frontend/FrontendOptions.h | 1 + lib/Driver/Types.cpp | 5 +++-- lib/Frontend/CompilerInvocation.cpp | 6 +++++- lib/Frontend/FrontendActions.cpp | 1 + lib/Frontend/FrontendOptions.cpp | 1 + test/Driver/cuda-simple.cu | 23 +++++++++++++++++++++++ test/Driver/lit.local.cfg | 3 ++- 8 files changed, 38 insertions(+), 5 deletions(-) create mode 100644 test/Driver/cuda-simple.cu diff --git a/include/clang/Driver/Types.def b/include/clang/Driver/Types.def index 32096792ca..adc12d34de 100644 --- a/include/clang/Driver/Types.def +++ b/include/clang/Driver/Types.def @@ -42,7 +42,8 @@ TYPE("cpp-output", PP_C, INVALID, "i", "u") TYPE("c", C, PP_C, "c", "u") TYPE("cl", CL, PP_C, "cl", "u") -TYPE("cuda", CUDA, PP_CXX, "cpp", "u") +TYPE("cuda-cpp-output", PP_CUDA, INVALID, "cui", "u") +TYPE("cuda", CUDA, PP_CUDA, "cu", "u") TYPE("objective-c-cpp-output", PP_ObjC, INVALID, "mi", "u") TYPE("objc-cpp-output", PP_ObjC_Alias, INVALID, "mi", "u") TYPE("objective-c", ObjC, PP_ObjC, "m", "u") diff --git a/include/clang/Frontend/FrontendOptions.h b/include/clang/Frontend/FrontendOptions.h index 71c5aa47af..c3aa226ea9 100644 --- a/include/clang/Frontend/FrontendOptions.h +++ b/include/clang/Frontend/FrontendOptions.h @@ -71,6 +71,7 @@ enum InputKind { IK_PreprocessedObjCXX, IK_OpenCL, IK_CUDA, + IK_PreprocessedCuda, IK_AST, IK_LLVM_IR }; diff --git a/lib/Driver/Types.cpp b/lib/Driver/Types.cpp index 6ee764c64e..7b28145755 100644 --- a/lib/Driver/Types.cpp +++ b/lib/Driver/Types.cpp @@ -85,7 +85,7 @@ bool types::isAcceptedByClang(ID Id) { case TY_Asm: case TY_C: case TY_PP_C: case TY_CL: - case TY_CUDA: + case TY_CUDA: case TY_PP_CUDA: case TY_ObjC: case TY_PP_ObjC: case TY_PP_ObjC_Alias: case TY_CXX: case TY_PP_CXX: case TY_ObjCXX: case TY_PP_ObjCXX: case TY_PP_ObjCXX_Alias: @@ -122,7 +122,7 @@ bool types::isCXX(ID Id) { case TY_ObjCXX: case TY_PP_ObjCXX: case TY_PP_ObjCXX_Alias: case TY_CXXHeader: case TY_PP_CXXHeader: case TY_ObjCXXHeader: case TY_PP_ObjCXXHeader: - case TY_CUDA: + case TY_CUDA: case TY_PP_CUDA: return true; } } @@ -153,6 +153,7 @@ types::ID types::lookupTypeForExtension(const char *Ext) { .Case("cl", TY_CL) .Case("cp", TY_CXX) .Case("cu", TY_CUDA) + .Case("cui", TY_PP_CUDA) .Case("hh", TY_CXXHeader) .Case("ll", TY_LLVM_IR) .Case("hpp", TY_CXXHeader) diff --git a/lib/Frontend/CompilerInvocation.cpp b/lib/Frontend/CompilerInvocation.cpp index 3c818b02fd..0b15935262 100644 --- a/lib/Frontend/CompilerInvocation.cpp +++ b/lib/Frontend/CompilerInvocation.cpp @@ -990,6 +990,7 @@ static InputKind ParseFrontendArgs(FrontendOptions &Opts, ArgList &Args, .Case("cpp-output", IK_PreprocessedC) .Case("assembler-with-cpp", IK_Asm) .Case("c++-cpp-output", IK_PreprocessedCXX) + .Case("cuda-cpp-output", IK_PreprocessedCuda) .Case("objective-c-cpp-output", IK_PreprocessedObjC) .Case("objc-cpp-output", IK_PreprocessedObjC) .Case("objective-c++-cpp-output", IK_PreprocessedObjCXX) @@ -1193,6 +1194,7 @@ void CompilerInvocation::setLangDefaults(LangOptions &Opts, InputKind IK, LangStd = LangStandard::lang_opencl; break; case IK_CUDA: + case IK_PreprocessedCuda: LangStd = LangStandard::lang_cuda; break; case IK_Asm: @@ -1245,7 +1247,8 @@ void CompilerInvocation::setLangDefaults(LangOptions &Opts, InputKind IK, Opts.NativeHalfType = 1; } - Opts.CUDA = LangStd == LangStandard::lang_cuda || IK == IK_CUDA; + Opts.CUDA = IK == IK_CUDA || IK == IK_PreprocessedCuda || + LangStd == LangStandard::lang_cuda; // OpenCL and C++ both have bool, true, false keywords. Opts.Bool = Opts.OpenCL || Opts.CPlusPlus; @@ -1360,6 +1363,7 @@ static void ParseLangArgs(LangOptions &Opts, ArgList &Args, InputKind IK, << A->getAsString(Args) << "OpenCL"; break; case IK_CUDA: + case IK_PreprocessedCuda: if (!Std.isCPlusPlus()) Diags.Report(diag::err_drv_argument_not_allowed_with) << A->getAsString(Args) << "CUDA"; diff --git a/lib/Frontend/FrontendActions.cpp b/lib/Frontend/FrontendActions.cpp index da08937b72..bc119297ae 100644 --- a/lib/Frontend/FrontendActions.cpp +++ b/lib/Frontend/FrontendActions.cpp @@ -690,6 +690,7 @@ void PrintPreambleAction::ExecuteAction() { case IK_None: case IK_Asm: case IK_PreprocessedC: + case IK_PreprocessedCuda: case IK_PreprocessedCXX: case IK_PreprocessedObjC: case IK_PreprocessedObjCXX: diff --git a/lib/Frontend/FrontendOptions.cpp b/lib/Frontend/FrontendOptions.cpp index 1869d0c78d..9ede674e47 100644 --- a/lib/Frontend/FrontendOptions.cpp +++ b/lib/Frontend/FrontendOptions.cpp @@ -18,6 +18,7 @@ InputKind FrontendOptions::getInputKindForExtension(StringRef Extension) { .Cases("S", "s", IK_Asm) .Case("i", IK_PreprocessedC) .Case("ii", IK_PreprocessedCXX) + .Case("cui", IK_PreprocessedCuda) .Case("m", IK_ObjC) .Case("mi", IK_PreprocessedObjC) .Cases("mm", "M", IK_ObjCXX) diff --git a/test/Driver/cuda-simple.cu b/test/Driver/cuda-simple.cu new file mode 100644 index 0000000000..99d4bfdc0e --- /dev/null +++ b/test/Driver/cuda-simple.cu @@ -0,0 +1,23 @@ +// Verify that we can parse a simple CUDA file with or without -save-temps +// http://llvm.org/PR22936 +// RUN: %clang -Werror -fsyntax-only -c %s +// +// Verify that we pass -x cuda-cpp-output to compiler after +// preprocessing a CUDA file +// RUN: %clang -Werror -### -save-temps -c %s 2>&1 | FileCheck %s +// CHECK: "-cc1" +// CHECK: "-E" +// CHECK: "-x" "cuda" +// CHECK-NEXT: "-cc1" +// CHECK: "-x" "cuda-cpp-output" +// +// Verify that compiler accepts CUDA syntax with "-x cuda-cpp-output". +// RUN: %clang -Werror -fsyntax-only -x cuda-cpp-output -c %s + +int cudaConfigureCall(int, int); +__attribute__((global)) void kernel() {} + +void func() { + kernel<<<1,1>>>(); +} + diff --git a/test/Driver/lit.local.cfg b/test/Driver/lit.local.cfg index d69e6ec131..af6d021390 100644 --- a/test/Driver/lit.local.cfg +++ b/test/Driver/lit.local.cfg @@ -1,4 +1,5 @@ -config.suffixes = ['.c', '.cpp', '.h', '.m', '.mm', '.S', '.s', '.f90', '.f95'] +config.suffixes = ['.c', '.cpp', '.h', '.m', '.mm', '.S', '.s', '.f90', '.f95', + '.cu'] config.substitutions = list(config.substitutions) config.substitutions.insert(0, ('%clang_cc1', -- 2.40.0