Skip to content

Commit 7b5d7e4

Browse files
authored
[SYCLomatic] Refine the migration of attribute __noinline__ to cover more scenarios (#2445)
Signed-off-by: intwanghao <[email protected]>
1 parent 9679fda commit 7b5d7e4

File tree

4 files changed

+15
-6
lines changed

4 files changed

+15
-6
lines changed

clang/lib/DPCT/ASTTraversal.cpp

Lines changed: 4 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -281,7 +281,10 @@ void IncludesCallbacks::MacroDefined(const Token &MacroNameTok,
281281

282282
// The "__noinline__" macro is re-defined and it is used in
283283
// "__attribute__()", do not migrate it.
284-
if (II->hasMacroDefinition() && (II->getName() == "__noinline__")) {
284+
if ((GetSourceFileType(
285+
DpctGlobalInfo::getInstance().getMainFile()->getFilePath()) ==
286+
SPT_CppSource) &&
287+
(II->getName() == "__noinline__")) {
285288
continue;
286289
}
287290

clang/lib/Lex/Preprocessor.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -818,7 +818,7 @@ bool Preprocessor::HandleIdentifier(Token &Identifier) {
818818
#ifdef SYCLomatic_CUSTOMIZATION
819819
// The "__noinline__" macro is re-defined and it is used in "__attribute__()",
820820
// do not handle it.
821-
if (LangOpts.CUDA && Identifier.is(tok::TokenKind::kw___noinline__) &&
821+
if ((II.getName() == "__noinline__") &&
822822
IsInAnalysisScopeFunc(Identifier.getLocation()) && IsInAttr) {
823823
return true;
824824
}

clang/test/dpct/cppnoinline.cpp

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,10 @@
1+
// RUN: dpct --format-range=none -out-root %T/cppnoinline %s --cuda-include-path="%cuda-path/include" -- -xc++
2+
// RUN: FileCheck %s --match-full-lines --input-file %T/cppnoinline/cppnoinline.cpp.dp.cpp
3+
// RUN: %if build_lit %{icpx -c -fsycl %T/cppnoinline/cppnoinline.cpp.dp.cpp -o %T/cppnoinline/cppnoinline.cpp.dp.o %}
4+
5+
#include <cuda_runtime.h>
6+
7+
// CHECK: #define NOINLINE __attribute__((__noinline__))
8+
#define NOINLINE __attribute__((__noinline__))
9+
// CHECK: NOINLINE void macro_in_macro_with_attr() {}
10+
NOINLINE void macro_in_macro_with_attr() {}

clang/test/dpct/noinline.cu

Lines changed: 0 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -52,7 +52,3 @@ __noinline__ __host__ __device__ scalar_t calc_igammac(scalar_t a, scalar_t b) {
5252

5353
// CHECK: __attribute__((__noinline__)) void macro_with_attr() {}
5454
__attribute__((__noinline__)) void macro_with_attr() {}
55-
// CHECK: #define NOINLINE __attribute__((__noinline__))
56-
#define NOINLINE __attribute__((__noinline__))
57-
// CHECK: NOINLINE void macro_in_macro_with_attr() {}
58-
NOINLINE void macro_in_macro_with_attr() {}

0 commit comments

Comments
 (0)