ffrankies created this revision.
ffrankies added reviewers: alexfh, jdoerfert, hokein, aaron.ballman.
ffrankies added projects: clang-tools-extra, clang, LLVM.
Herald added subscribers: mgehre, arphaman, xazax.hun, Anastasia, mgorny.
This lint check is a part of the FLOCL (FPGA Linters for OpenCL) project out of
the Synergy Lab at Virginia Tech.
FLOCL is a set of lint checks aimed at FPGA developers who write code in OpenCL.
The altera ID dependent backward branch lint check finds ID dependent variables
and fields used within loops, and warns of their usage. Using these variables
in loops can lead to performance degradation.
Repository:
rCTE Clang Tools Extra
https://reviews.llvm.org/D70094
Files:
clang-tidy/CMakeLists.txt
clang-tidy/ClangTidyForceLinker.h
clang-tidy/altera/AlteraTidyModule.cpp
clang-tidy/altera/CMakeLists.txt
clang-tidy/altera/IdDependentBackwardBranchCheck.cpp
clang-tidy/altera/IdDependentBackwardBranchCheck.h
docs/ReleaseNotes.rst
docs/clang-tidy/checks/altera-id-dependent-backward-branch.rst
docs/clang-tidy/index.rst
test/clang-tidy/checkers/altera-id-dependent-backward-branch.cpp
Index: test/clang-tidy/checkers/altera-id-dependent-backward-branch.cpp
===================================================================
--- /dev/null
+++ test/clang-tidy/checkers/altera-id-dependent-backward-branch.cpp
@@ -0,0 +1,83 @@
+// RUN: %check_clang_tidy %s altera-id-dependent-backward-branch %t -- -header-filter=.* "--" -cl-std=CL1.2 -c --include opencl-c.h
+
+typedef struct ExampleStruct {
+ int IDDepField;
+} ExampleStruct;
+
+void error() {
+ // ==== Assignments ====
+ int ThreadID = get_local_id(0);
+// CHECK-NOTES: :[[@LINE-1]]:3: warning: assignment of ID-dependent variable ThreadID [altera-id-dependent-backward-branch]
+
+ ExampleStruct Example;
+ Example.IDDepField = get_local_id(0);
+// CHECK-NOTES: :[[@LINE-1]]:3: warning: assignment of ID-dependent field IDDepField [altera-id-dependent-backward-branch]
+
+ // ==== Inferred Assignments ====
+ int ThreadID2 = ThreadID * get_local_size(0);
+// CHECK-NOTES: :[[@LINE-1]]:3: warning: inferred assignment of ID-dependent value from ID-dependent variable ThreadID [altera-id-dependent-backward-branch]
+
+ int ThreadID3 = Example.IDDepField; // OK: not used in any loops
+
+ ExampleStruct UnusedStruct = {
+ ThreadID * 2 // OK: not used in any loops
+ };
+
+ // ==== Conditional Expressions ====
+ int accumulator = 0;
+ for (int i = 0; i < get_local_id(0); i++) {
+ // CHECK-NOTES: :[[@LINE-1]]:19: warning: backward branch (for loop) is ID-dependent due to ID function call and may cause performance degradation [altera-id-dependent-backward-branch]
+ accumulator++;
+ }
+
+ int j = 0;
+ while (j < get_local_id(0)) {
+ // CHECK-NOTES: :[[@LINE-1]]:10: warning: backward branch (while loop) is ID-dependent due to ID function call and may cause performance degradation [altera-id-dependent-backward-branch]
+ accumulator++;
+ }
+
+ do {
+ accumulator++;
+ } while (j < get_local_id(0));
+ // CHECK-NOTES: :[[@LINE-1]]:12: warning: backward branch (do loop) is ID-dependent due to ID function call and may cause performance degradation [altera-id-dependent-backward-branch]
+
+ for (int i = 0; i < ThreadID2; i++) {
+ // CHECK-NOTES: :[[@LINE-1]]:19: warning: backward branch (for loop) is ID-dependent due to variable reference to 'ThreadID2' and may cause performance degradation [altera-id-dependent-backward-branch]
+ accumulator++;
+ }
+
+ while (j < ThreadID) {
+ // CHECK-NOTES: :[[@LINE-1]]:10: warning: backward branch (while loop) is ID-dependent due to variable reference to 'ThreadID' and may cause performance degradation [altera-id-dependent-backward-branch]
+ accumulator++;
+ }
+
+ do {
+ accumulator++;
+ } while (j < ThreadID);
+ // CHECK-NOTES: :[[@LINE-1]]:12: warning: backward branch (do loop) is ID-dependent due to variable reference to 'ThreadID' and may cause performance degradation [altera-id-dependent-backward-branch]
+
+ for (int i = 0; i < Example.IDDepField; i++) {
+ // CHECK-NOTES: :[[@LINE-1]]:19: warning: backward branch (for loop) is ID-dependent due to member reference to 'IDDepField' and may cause performance degradation [altera-id-dependent-backward-branch]
+ accumulator++;
+ }
+
+ while (j < Example.IDDepField) {
+ // CHECK-NOTES: :[[@LINE-1]]:10: warning: backward branch (while loop) is ID-dependent due to member reference to 'IDDepField' and may cause performance degradation [altera-id-dependent-backward-branch]
+ accumulator++;
+ }
+
+ do {
+ accumulator++;
+ } while (j < Example.IDDepField);
+ // CHECK-NOTES: :[[@LINE-1]]:12: warning: backward branch (do loop) is ID-dependent due to member reference to 'IDDepField' and may cause performance degradation [altera-id-dependent-backward-branch]
+}
+
+void success() {
+ int accumulator = 0;
+
+ for (int i = 0; i < 1000; i++) {
+ if (i < get_local_id(0)) {
+ accumulator++;
+ }
+ }
+}
Index: docs/clang-tidy/index.rst
===================================================================
--- docs/clang-tidy/index.rst
+++ docs/clang-tidy/index.rst
@@ -59,6 +59,7 @@
====================== =========================================================
``abseil-`` Checks related to Abseil library.
``android-`` Checks related to Android.
+``altera-`` Checks related to OpenCL programming for FPGAs.
``boost-`` Checks related to Boost library.
``bugprone-`` Checks that target bugprone code constructs.
``cert-`` Checks related to CERT Secure Coding Guidelines.
Index: docs/clang-tidy/checks/altera-id-dependent-backward-branch.rst
===================================================================
--- /dev/null
+++ docs/clang-tidy/checks/altera-id-dependent-backward-branch.rst
@@ -0,0 +1,29 @@
+.. title:: clang-tidy - fpga-id-dependent-backward-branch
+
+fpga-id-dependent-backward-branch
+=================================
+
+Finds ID-dependent variables and fields that are used within loops. This causes
+branches to occur inside the loops, and thus leads to performance degradation.
+
+Based on the `Altera SDK for OpenCL: Best Practices Guide
+<https://www.altera.com/en_US/pdfs/literature/hb/opencl-sdk/aocl_optimization_guide.pdf>`_.
+
+.. code-block:: c++
+
+ // The following code will produce a warning because this ID-dependent
+ // variable is used in a loop condition statement.
+ int ThreadID = get_local_id(0);
+
+ // The following loop will produce a warning because the loop condition
+ // statement depends on an ID-dependent variable.
+ for (int i = 0; i < ThreadID; ++i) {
+ std::cout << i << std::endl;
+ }
+
+ // The following loop will not produce a warning, because the ID-dependent
+ // variable is not used in the loop condition statement.
+ for (int i = 0; i < 100; ++i) {
+ std::cout << ThreadID << std::endl;
+ }
+
Index: docs/ReleaseNotes.rst
===================================================================
--- docs/ReleaseNotes.rst
+++ docs/ReleaseNotes.rst
@@ -67,6 +67,17 @@
Improvements to clang-tidy
--------------------------
+- New :doc:`altera <clang-tidy/modules/altera>` module.
+
+ Includes checks related to OpenCL for FPGA coding guidelines, based on the
+ `Altera SDK for OpenCL: Best Practices Guide
+ <https://www.altera.com/en_US/pdfs/literature/hb/opencl-sdk/aocl_optimization_guide.pdf>`_.
+
+- New :doc:`altera-id-dependent-backward-branch
+ <clang-tidy/checks/altera-id-dependent-backward-branch>` check.
+
+ Finds ID-dependent variables and fields that are used within loops.
+
- New :doc:`bugprone-dynamic-static-initializers
<clang-tidy/checks/bugprone-dynamic-static-initializers>` check.
Index: clang-tidy/altera/IdDependentBackwardBranchCheck.h
===================================================================
--- /dev/null
+++ clang-tidy/altera/IdDependentBackwardBranchCheck.h
@@ -0,0 +1,82 @@
+//===--- IdDependentBackwardBranchCheck.h - clang-tidy-----------*- C++ -*-===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#ifndef LLVM_CLANG_TOOLS_EXTRA_CLANG_TIDY_ALTERA_ID_DEPENDENT_BACKWARD_BRANCH_H
+#define LLVM_CLANG_TOOLS_EXTRA_CLANG_TIDY_ALTERA_ID_DEPENDENT_BACKWARD_BRANCH_H
+
+#include "../ClangTidy.h"
+
+namespace clang {
+namespace tidy {
+namespace altera {
+
+/// Finds ID-dependent variables and fields used within loops, and warns of
+/// their usage. Using these variables in loops can lead to performance
+/// degradation.
+///
+/// For the user-facing documentation see:
+/// http://clang.llvm.org/extra/clang-tidy/checks/altera-id-dependent-backward-branch.html
+class IdDependentBackwardBranchCheck : public ClangTidyCheck {
+private:
+ enum LoopType { UNK_LOOP = -1, DO_LOOP = 0, WHILE_LOOP = 1, FOR_LOOP = 2 };
+ // Stores information necessary for printing out source of error.
+ struct IdDependencyRecord {
+ IdDependencyRecord(const VarDecl *Declaration, SourceLocation Location,
+ std::string Message)
+ : VariableDeclaration(Declaration), Location(Location),
+ Message(Message) {}
+ IdDependencyRecord(const FieldDecl *Declaration, SourceLocation Location,
+ std::string Message)
+ : FieldDeclaration(Declaration), Location(Location), Message(Message) {}
+ IdDependencyRecord() {}
+ const VarDecl *VariableDeclaration;
+ const FieldDecl *FieldDeclaration;
+ SourceLocation Location;
+ std::string Message;
+ };
+ // Stores the locations where ID-dependent variables are created.
+ std::map<const VarDecl *, IdDependencyRecord> IdDepVarsMap;
+ // Stores the locations where ID-dependent fields are created.
+ std::map<const FieldDecl *, IdDependencyRecord> IdDepFieldsMap;
+ /// Returns an IdDependencyRecord if the Expression contains an ID-dependent
+ /// variable, returns a nullptr otherwise.
+ IdDependencyRecord *hasIdDepVar(const Expr *Expression);
+ /// Returns an IdDependencyRecord if the Expression contains an ID-dependent
+ /// field, returns a nullptr otherwise.
+ IdDependencyRecord *hasIdDepField(const Expr *Expression);
+ /// Stores the location an ID-dependent variable is created from a call to
+ /// an ID function in IdDepVarsMap.
+ void saveIdDepVar(const Stmt *Statement, const VarDecl *Variable);
+ /// Stores the location an ID-dependent field is created from a call to an ID
+ /// function in IdDepFieldsMap.
+ void saveIdDepField(const Stmt *Statement, const FieldDecl *Field);
+ /// Stores the location an ID-dependent variable is created from a reference
+ /// to another ID-dependent variable or field in IdDepVarsMap.
+ void saveIdDepVarFromReference(const DeclRefExpr *RefExpr,
+ const MemberExpr *MemExpr,
+ const VarDecl *PotentialVar);
+ /// Stores the location an ID-dependent field is created from a reference to
+ /// another ID-dependent variable or field in IdDepFieldsMap.
+ void saveIdDepFieldFromReference(const DeclRefExpr *RefExpr,
+ const MemberExpr *MemExpr,
+ const FieldDecl *PotentialField);
+ /// Returns the loop type.
+ LoopType getLoopType(const Stmt *Loop);
+
+public:
+ IdDependentBackwardBranchCheck(StringRef Name, ClangTidyContext *Context)
+ : ClangTidyCheck(Name, Context) {}
+ void registerMatchers(ast_matchers::MatchFinder *Finder) override;
+ void check(const ast_matchers::MatchFinder::MatchResult &Result) override;
+};
+
+} // namespace altera
+} // namespace tidy
+} // namespace clang
+
+#endif // LLVM_CLANG_TOOLS_EXTRA_CLANG_TIDY_ALTERA_ID_DEPENDENT_BACKWARD_BRANCH_H
Index: clang-tidy/altera/IdDependentBackwardBranchCheck.cpp
===================================================================
--- /dev/null
+++ clang-tidy/altera/IdDependentBackwardBranchCheck.cpp
@@ -0,0 +1,295 @@
+//===--- IdDependentBackwardBranchCheck.cpp - clang-tidy ------------------===//
+//
+// The LLVM Compiler Infrastructure
+//
+// This file is distributed under the University of Illinois Open Source
+// License. See LICENSE.TXT for details.
+//
+//===----------------------------------------------------------------------===//
+
+#include "IdDependentBackwardBranchCheck.h"
+#include "clang/AST/ASTContext.h"
+#include "clang/ASTMatchers/ASTMatchFinder.h"
+#include <sstream>
+#include <vector>
+
+using namespace clang::ast_matchers;
+
+namespace clang {
+namespace tidy {
+namespace altera {
+
+void IdDependentBackwardBranchCheck::registerMatchers(MatchFinder *Finder) {
+ // Prototype to identify all variables which hold a thread-variant ID
+ // First Matcher just finds all the direct assignments of either ID call
+ const auto TID_RHS = expr(hasDescendant(callExpr(callee(functionDecl(
+ anyOf(hasName("get_global_id"), hasName("get_local_id")))))));
+
+ const auto ANY_ASSIGN = anyOf(
+ hasOperatorName("="), hasOperatorName("*="), hasOperatorName("/="),
+ hasOperatorName("%="), hasOperatorName("+="), hasOperatorName("-="),
+ hasOperatorName("<<="), hasOperatorName(">>="), hasOperatorName("&="),
+ hasOperatorName("^="), hasOperatorName("|="));
+
+ Finder->addMatcher(
+ compoundStmt(
+ // Bind on actual get_local/global_id calls
+ forEachDescendant(
+ stmt(anyOf(declStmt(hasDescendant(varDecl(hasInitializer(TID_RHS))
+ .bind("tid_dep_var"))),
+ binaryOperator(allOf(
+ ANY_ASSIGN, hasRHS(TID_RHS),
+ hasLHS(anyOf(
+ declRefExpr(to(varDecl().bind("tid_dep_var"))),
+ memberExpr(member(
+ fieldDecl().bind("tid_dep_field")))))))))
+ .bind("straight_assignment"))),
+ this);
+
+ // Bind all VarDecls that include an initializer with a variable DeclRefExpr
+ // (incase it is ID-dependent)
+ Finder->addMatcher(
+ stmt(forEachDescendant(
+ varDecl(
+ hasInitializer(forEachDescendant(stmt(anyOf(
+ declRefExpr(to(varDecl())).bind("assign_ref_var"),
+ memberExpr(member(fieldDecl())).bind("assign_ref_field"))))))
+ .bind("pot_tid_var"))),
+ this);
+
+ // Bind all VarDecls that are assigned a value with a variable DeclRefExpr (in
+ // case it is ID-dependent)
+ Finder->addMatcher(
+ stmt(forEachDescendant(binaryOperator(allOf(
+ ANY_ASSIGN,
+ hasRHS(forEachDescendant(stmt(anyOf(
+ declRefExpr(to(varDecl())).bind("assign_ref_var"),
+ memberExpr(member(fieldDecl())).bind("assign_ref_field"))))),
+ hasLHS(
+ anyOf(declRefExpr(to(varDecl().bind("pot_tid_var"))),
+ memberExpr(member(fieldDecl().bind("pot_tid_field"))))))))),
+ this);
+
+ // Second Matcher looks for branch statements inside of loops and bind on the
+ // condition expression IF it either calls an ID function or has a variable
+ // DeclRefExpr DeclRefExprs are checked later to confirm whether the variable
+ // is ID-dependent
+ const auto COND_EXPR =
+ expr(anyOf(hasDescendant(callExpr(callee(functionDecl(
+ anyOf(hasName("get_global_id"),
+ hasName("get_local_id")))))
+ .bind("id_call")),
+ hasDescendant(stmt(anyOf(declRefExpr(to(varDecl())),
+ memberExpr(member(fieldDecl())))))))
+ .bind("cond_expr");
+ Finder->addMatcher(stmt(anyOf(forStmt(hasCondition(COND_EXPR)),
+ doStmt(hasCondition(COND_EXPR)),
+ whileStmt(hasCondition(COND_EXPR))))
+ .bind("backward_branch"),
+ this);
+}
+
+IdDependentBackwardBranchCheck::IdDependencyRecord *
+IdDependentBackwardBranchCheck::hasIdDepVar(const Expr *Expression) {
+ if (const DeclRefExpr *expr = dyn_cast<DeclRefExpr>(Expression)) {
+ // It is a DeclRefExpr, so check if it's an ID-dependent variable
+ const VarDecl *CheckVariable = dyn_cast<VarDecl>(expr->getDecl());
+ auto FoundVariable = IdDepVarsMap.find(CheckVariable);
+ if (FoundVariable == IdDepVarsMap.end()) {
+ return nullptr;
+ }
+ return &(FoundVariable->second);
+ }
+ for (auto i = Expression->child_begin(), e = Expression->child_end(); i != e;
+ ++i) {
+ if (auto *ChildExpression = dyn_cast<Expr>(*i)) {
+ auto Result = hasIdDepVar(ChildExpression);
+ if (Result) {
+ return Result;
+ }
+ }
+ }
+ return nullptr;
+}
+
+IdDependentBackwardBranchCheck::IdDependencyRecord *
+IdDependentBackwardBranchCheck::hasIdDepField(const Expr *Expression) {
+ if (const MemberExpr *MemberExpression = dyn_cast<MemberExpr>(Expression)) {
+ const FieldDecl *CheckField =
+ dyn_cast<FieldDecl>(MemberExpression->getMemberDecl());
+ auto FoundField = IdDepFieldsMap.find(CheckField);
+ if (FoundField == IdDepFieldsMap.end()) {
+ return nullptr;
+ }
+ return &(FoundField->second);
+ }
+ for (auto I = Expression->child_begin(), E = Expression->child_end(); I != E;
+ ++I) {
+ if (auto *ChildExpression = dyn_cast<Expr>(*I)) {
+ auto Result = hasIdDepField(ChildExpression);
+ if (Result) {
+ return Result;
+ }
+ }
+ }
+ return nullptr;
+}
+
+void IdDependentBackwardBranchCheck::saveIdDepVar(const Stmt *Statement,
+ const VarDecl *Variable) {
+ // Record that this variable is thread-dependent
+ std::ostringstream StringStream;
+ StringStream << "assignment of ID-dependent variable "
+ << Variable->getNameAsString();
+ IdDepVarsMap[Variable] =
+ IdDependencyRecord(Variable, Variable->getBeginLoc(), StringStream.str());
+}
+
+void IdDependentBackwardBranchCheck::saveIdDepField(const Stmt *Statement,
+ const FieldDecl *Field) {
+ std::ostringstream StringStream;
+ StringStream << "assignment of ID-dependent field "
+ << Field->getNameAsString();
+ IdDepFieldsMap[Field] =
+ IdDependencyRecord(Field, Statement->getBeginLoc(), StringStream.str());
+}
+
+void IdDependentBackwardBranchCheck::saveIdDepVarFromReference(
+ const DeclRefExpr *RefExpr, const MemberExpr *MemExpr,
+ const VarDecl *PotentialVar) {
+ // If the variable is already in IdDepVarsMap, ignore it
+ if (IdDepVarsMap.find(PotentialVar) != IdDepVarsMap.end()) {
+ return;
+ }
+ std::ostringstream StringStream;
+ StringStream << "inferred assignment of ID-dependent value from "
+ "ID-dependent ";
+ if (RefExpr) {
+ const auto RefVar = dyn_cast<VarDecl>(RefExpr->getDecl());
+ // If variable isn't ID-dependent, but refVar is
+ if (IdDepVarsMap.find(RefVar) != IdDepVarsMap.end()) {
+ StringStream << "variable " << RefVar->getNameAsString();
+ }
+ }
+ if (MemExpr) {
+ const auto RefField = dyn_cast<FieldDecl>(MemExpr->getMemberDecl());
+ // If variable isn't ID-dependent, but refField is
+ if (IdDepFieldsMap.find(RefField) != IdDepFieldsMap.end()) {
+ StringStream << "member " << RefField->getNameAsString();
+ }
+ }
+ IdDepVarsMap[PotentialVar] = IdDependencyRecord(
+ PotentialVar, PotentialVar->getBeginLoc(), StringStream.str());
+}
+
+void IdDependentBackwardBranchCheck::saveIdDepFieldFromReference(
+ const DeclRefExpr *RefExpr, const MemberExpr *MemExpr,
+ const FieldDecl *PotentialField) {
+ // If the field is already in IdDepFieldsMap, ignore it
+ if (IdDepFieldsMap.find(PotentialField) != IdDepFieldsMap.end()) {
+ return;
+ }
+ std::ostringstream StringStream;
+ StringStream << "inferred assignment of ID-dependent member from "
+ "ID-dependent ";
+ if (RefExpr) {
+ const auto RefVar = dyn_cast<VarDecl>(RefExpr->getDecl());
+ // If field isn't ID-dependent, but RefVar is
+ if (IdDepVarsMap.find(RefVar) != IdDepVarsMap.end()) {
+ StringStream << "variable " << RefVar->getNameAsString();
+ }
+ }
+ if (MemExpr) {
+ const auto RefField = dyn_cast<FieldDecl>(MemExpr->getMemberDecl());
+ if (IdDepFieldsMap.find(RefField) != IdDepFieldsMap.end()) {
+ StringStream << "member " << RefField->getNameAsString();
+ }
+ }
+ IdDepFieldsMap[PotentialField] = IdDependencyRecord(
+ PotentialField, PotentialField->getBeginLoc(), StringStream.str());
+}
+
+IdDependentBackwardBranchCheck::LoopType
+IdDependentBackwardBranchCheck::getLoopType(const Stmt *Loop) {
+ if (const auto DoLoop = dyn_cast<DoStmt>(Loop)) {
+ return DO_LOOP; // loop_type = 0;
+ } else if (const auto WhileLoop = dyn_cast<WhileStmt>(Loop)) {
+ return WHILE_LOOP; // loop_type = 1;
+ } else if (const auto ForLoop = dyn_cast<ForStmt>(Loop)) {
+ return FOR_LOOP; // loop_type = 2;
+ }
+ return UNK_LOOP;
+}
+
+void IdDependentBackwardBranchCheck::check(
+ const MatchFinder::MatchResult &Result) {
+ // The first half of the callback only deals with identifying and propagating
+ // ID-dependency information into the IdDepVars vector
+ const auto *Variable = Result.Nodes.getNodeAs<VarDecl>("tid_dep_var");
+ const auto *Field = Result.Nodes.getNodeAs<FieldDecl>("tid_dep_field");
+ const auto *Statement = Result.Nodes.getNodeAs<Stmt>("straight_assignment");
+ const auto *RefExpr = Result.Nodes.getNodeAs<DeclRefExpr>("assign_ref_var");
+ const auto *MemExpr = Result.Nodes.getNodeAs<MemberExpr>("assign_ref_field");
+ const auto *PotentialVar = Result.Nodes.getNodeAs<VarDecl>("pot_tid_var");
+ const auto *PotentialField =
+ Result.Nodes.getNodeAs<FieldDecl>("pot_tid_field");
+
+ // Save variables and fields assigned directly through ID function calls
+ if (Statement && (Variable || Field)) {
+ if (Variable) {
+ saveIdDepVar(Statement, Variable);
+ } else if (Field) {
+ saveIdDepField(Statement, Field);
+ }
+ }
+
+ // Save variables assigned to values of Id-dependent variables and fields
+ if ((RefExpr || MemExpr) && PotentialVar) {
+ saveIdDepVarFromReference(RefExpr, MemExpr, PotentialVar);
+ }
+
+ // Save fields assigned to values of ID-dependent variables and fields
+ if ((RefExpr || MemExpr) && PotentialField) {
+ saveIdDepFieldFromReference(RefExpr, MemExpr, PotentialField);
+ }
+
+ // The second part of the callback deals with checking if a branch inside a
+ // loop is thread dependent
+ const auto *CondExpr = Result.Nodes.getNodeAs<Expr>("cond_expr");
+ const auto *IDCall = Result.Nodes.getNodeAs<CallExpr>("id_call");
+ const auto *Loop = Result.Nodes.getNodeAs<Stmt>("backward_branch");
+ if (!Loop) {
+ return;
+ }
+ LoopType Type = getLoopType(Loop);
+ if (CondExpr) {
+ if (IDCall) { // If conditional expression calls an ID function directly
+ diag(CondExpr->getBeginLoc(),
+ "backward branch (%select{do|while|for}0 loop) is ID-dependent due "
+ "to ID function call and may cause performance degradation")
+ << Type;
+ } else { // Conditional expression has DeclRefExpr(s), check ID-dependency
+ IdDependencyRecord *IdDepVar = hasIdDepVar(CondExpr);
+ IdDependencyRecord *IdDepField = hasIdDepField(CondExpr);
+ if (IdDepVar) {
+ diag(IdDepVar->Location, IdDepVar->Message);
+ diag(
+ CondExpr->getBeginLoc(),
+ "backward branch (%select{do|while|for}0 loop) is ID-dependent due "
+ "to variable reference to %1 and may cause performance degradation")
+ << Type << IdDepVar->VariableDeclaration;
+ } else if (IdDepField) {
+ diag(IdDepField->Location, IdDepField->Message);
+ diag(
+ CondExpr->getBeginLoc(),
+ "backward branch (%select{do|while|for}0 loop) is ID-dependent due "
+ "to member reference to %1 and may cause performance degradation")
+ << Type << IdDepField->FieldDeclaration;
+ }
+ }
+ }
+}
+
+} // namespace altera
+} // namespace tidy
+} // namespace clang
Index: clang-tidy/altera/CMakeLists.txt
===================================================================
--- /dev/null
+++ clang-tidy/altera/CMakeLists.txt
@@ -0,0 +1,15 @@
+set(LLVM_LINK_COMPONENTS support)
+
+add_clang_library(clangTidyAlteraModule
+ AlteraTidyModule.cpp
+ IdDependentBackwardBranchCheck.cpp
+
+ LINK_LIBS
+ clangAnalysis
+ clangAST
+ clangASTMatchers
+ clangBasic
+ clangLex
+ clangTidy
+ clangTidyUtils
+ )
Index: clang-tidy/altera/AlteraTidyModule.cpp
===================================================================
--- /dev/null
+++ clang-tidy/altera/AlteraTidyModule.cpp
@@ -0,0 +1,39 @@
+//===--- AlteraTidyModule.cpp - clang-tidy --------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include "../ClangTidy.h"
+#include "../ClangTidyModule.h"
+#include "../ClangTidyModuleRegistry.h"
+#include "IdDependentBackwardBranchCheck.h"
+
+using namespace clang::ast_matchers;
+
+namespace clang {
+namespace tidy {
+namespace altera {
+
+class AlteraModule : public ClangTidyModule {
+public:
+ void addCheckFactories(ClangTidyCheckFactories &CheckFactories) override {
+ CheckFactories.registerCheck<IdDependentBackwardBranchCheck>(
+ "altera-id-dependent-backward-branch");
+ }
+};
+
+} // namespace altera
+
+// Register the AlteraTidyModule using this statically initialized variable.
+static ClangTidyModuleRegistry::Add<altera::AlteraModule>
+ X("altera-module", "Adds Altera FPGA OpenCL lint checks.");
+
+// This anchor is used to force the linker to link in the generated object file
+// and thus register the AlteraModule.
+volatile int AlteraModuleAnchorSource = 0;
+
+} // namespace tidy
+} // namespace clang
Index: clang-tidy/ClangTidyForceLinker.h
===================================================================
--- clang-tidy/ClangTidyForceLinker.h
+++ clang-tidy/ClangTidyForceLinker.h
@@ -25,6 +25,11 @@
static int LLVM_ATTRIBUTE_UNUSED AbseilModuleAnchorDestination =
AbseilModuleAnchorSource;
+// This anchor is used to force the linker to link the AlteraModule.
+extern volatile int AlteraModuleAnchorSource;
+static int LLVM_ATTRIBUTE_UNUSED AlteraModuleAnchorDestination =
+ AlteraModuleAnchorSource;
+
// This anchor is used to force the linker to link the BoostModule.
extern volatile int BoostModuleAnchorSource;
static int LLVM_ATTRIBUTE_UNUSED BoostModuleAnchorDestination =
Index: clang-tidy/CMakeLists.txt
===================================================================
--- clang-tidy/CMakeLists.txt
+++ clang-tidy/CMakeLists.txt
@@ -41,6 +41,7 @@
# If you add a check, also add it to ClangTidyForceLinker.h in this directory.
add_subdirectory(android)
add_subdirectory(abseil)
+add_subdirectory(altera)
add_subdirectory(boost)
add_subdirectory(bugprone)
add_subdirectory(cert)
@@ -65,6 +66,7 @@
set(ALL_CLANG_TIDY_CHECKS
clangTidyAndroidModule
clangTidyAbseilModule
+ clangTidyAlteraModule
clangTidyBoostModule
clangTidyBugproneModule
clangTidyCERTModule
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
