[NFC][Clang][OpenMP] Add helper functions/utils for finding/comparing attach base-ptrs. #155625
+566
−0
Conversation
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
…base-ptrs. These have been pulled out of the codegen PR llvm#153683, to reduce the size of that PR.
llvmbot
added
clang
|
@llvm/pr-subscribers-clang-codegen Author: Abhinav Gaba (abhinavgaba) ChangesThese have been pulled out of the codegen PR #153683, to reduce the size of that PR. Patch is 25.43 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/155625.diff 5 Files Affected:
diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h
index 1118d3e062e68..9627e99a306b4 100644
--- a/clang/include/clang/AST/OpenMPClause.h
+++ b/clang/include/clang/AST/OpenMPClause.h
@@ -5815,6 +5815,12 @@ class OMPClauseMappableExprCommon {
ValueDecl *getAssociatedDeclaration() const {
return AssociatedDeclaration;
}
+
+ bool operator==(const MappableComponent &Other) const {
+ return AssociatedExpressionNonContiguousPr ==
+ Other.AssociatedExpressionNonContiguousPr &&
+ AssociatedDeclaration == Other.AssociatedDeclaration;
+ }
};
// List of components of an expression. This first one is the whole
@@ -5828,6 +5834,95 @@ class OMPClauseMappableExprCommon {
using MappableExprComponentLists = SmallVector<MappableExprComponentList, 8>;
using MappableExprComponentListsRef = ArrayRef<MappableExprComponentList>;
+ // Hash function to allow usage as DenseMap keys.
+ friend llvm::hash_code hash_value(const MappableComponent &MC) {
+ return llvm::hash_combine(MC.getAssociatedExpression(),
+ MC.getAssociatedDeclaration(),
+ MC.isNonContiguous());
+ }
+
+public:
+ /// Get the type of an element of a ComponentList Expr \p Exp.
+ ///
+ /// For something like the following:
+ /// ```c
+ /// int *p, **p;
+ /// ```
+ /// The types for the following Exprs would be:
+ /// Expr | Type
+ /// ---------|-----------
+ /// p | int *
+ /// *p | int
+ /// p[0] | int
+ /// p[0:1] | int
+ /// pp | int **
+ /// pp[0] | int *
+ /// pp[0:1] | int *
+ /// Note: this assumes that if \p Exp is an array-section, it is contiguous.
+ static QualType getComponentExprElementType(const Expr *Exp);
+
+ /// Find the attach pointer expression from a list of mappable expression
+ /// components.
+ ///
+ /// This function traverses the component list to find the first
+ /// expression that has a pointer type, which represents the attach
+ /// base pointer expr for the current component-list.
+ ///
+ /// For example, given the following:
+ ///
+ /// ```c
+ /// struct S {
+ /// int a;
+ /// int b[10];
+ /// int c[10][10];
+ /// int *p;
+ /// int **pp;
+ /// }
+ /// S s, *ps, **pps, *(pas[10]), ***ppps;
+ /// int i;
+ /// ```
+ ///
+ /// The base-pointers for the following map operands would be:
+ /// map list-item | attach base-pointer | attach base-pointer
+ /// | for directives except | target_update (if
+ /// | target_update | different)
+ /// ----------------|-----------------------|---------------------
+ /// s | N/A |
+ /// s.a | N/A |
+ /// s.p | N/A |
+ /// ps | N/A |
+ /// ps->p | ps |
+ /// ps[1] | ps |
+ /// *(ps + 1) | ps |
+ /// (ps + 1)[1] | ps |
+ /// ps[1:10] | ps |
+ /// ps->b[10] | ps |
+ /// ps->p[10] | ps->p |
+ /// ps->c[1][2] | ps |
+ /// ps->c[1:2][2] | (error diagnostic) | N/A, TODO: ps
+ /// ps->c[1:1][2] | ps | N/A, TODO: ps
+ /// pps[1][2] | pps[1] |
+ /// pps[1:1][2] | pps[1:1] | N/A, TODO: pps[1:1]
+ /// pps[1:i][2] | pps[1:i] | N/A, TODO: pps[1:i]
+ /// pps[1:2][2] | (error diagnostic) | N/A
+ /// pps[1]->p | pps[1] |
+ /// pps[1]->p[10] | pps[1] |
+ /// pas[1] | N/A |
+ /// pas[1][2] | pas[1] |
+ /// ppps[1][2] | ppps[1] |
+ /// ppps[1][2][3] | ppps[1][2] |
+ /// ppps[1][2:1][3] | ppps[1][2:1] | N/A, TODO: ppps[1][2:1]
+ /// ppps[1][2:2][3] | (error diagnostic) | N/A
+ /// Returns a pair of the attach pointer expression and its depth in the
+ /// component list.
+ /// TODO: This may need to be updated to handle ref_ptr/ptee cases for byref
+ /// map operands.
+ /// TODO: Handle cases for target-update, where the list-item is a
+ /// non-contiguous array-section that still has a base-pointer.
+ static std::pair<const Expr *, std::optional<size_t>>
+ findAttachPtrExpr(MappableExprComponentListRef Components,
+ OpenMPDirectiveKind CurDirKind);
+
protected:
// Return the total number of elements in a list of component lists.
static unsigned
diff --git a/clang/include/clang/Basic/OpenMPKinds.h b/clang/include/clang/Basic/OpenMPKinds.h
index f40db4c13c55a..e37887e8b86ba 100644
--- a/clang/include/clang/Basic/OpenMPKinds.h
+++ b/clang/include/clang/Basic/OpenMPKinds.h
@@ -301,6 +301,14 @@ bool isOpenMPTargetExecutionDirective(OpenMPDirectiveKind DKind);
/// otherwise - false.
bool isOpenMPTargetDataManagementDirective(OpenMPDirectiveKind DKind);
+/// Checks if the specified directive is a map-entering target directive.
+/// \param DKind Specified directive.
+/// \return true - the directive is a map-entering target directive like
+/// 'omp target', 'omp target data', 'omp target enter data',
+/// 'omp target parallel', etc. (excludes 'omp target exit data', 'omp target
+/// update') otherwise - false.
+bool isOpenMPTargetMapEnteringDirective(OpenMPDirectiveKind DKind);
+
/// Checks if the specified composite/combined directive constitutes a teams
/// directive in the outermost nest. For example
/// 'omp teams distribute' or 'omp teams distribute parallel for'.
diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp
index 588b0dcc6d7b8..eff897a1a33b2 100644
--- a/clang/lib/AST/OpenMPClause.cpp
+++ b/clang/lib/AST/OpenMPClause.cpp
@@ -15,6 +15,7 @@
#include "clang/AST/Attr.h"
#include "clang/AST/Decl.h"
#include "clang/AST/DeclOpenMP.h"
+#include "clang/AST/ExprOpenMP.h"
#include "clang/Basic/LLVM.h"
#include "clang/Basic/OpenMPKinds.h"
#include "clang/Basic/TargetInfo.h"
@@ -1156,6 +1157,73 @@ unsigned OMPClauseMappableExprCommon::getUniqueDeclarationsTotalNumber(
return UniqueDecls.size();
}
+QualType
+OMPClauseMappableExprCommon::getComponentExprElementType(const Expr *Exp) {
+ assert(!isa<OMPArrayShapingExpr>(Exp) &&
+ "Cannot get element-type from array-shaping expr.");
+
+ // Unless we are handling array-section expressions, including
+ // array-subscripts, derefs, we can rely on getType.
+ if (!isa<ArraySectionExpr>(Exp))
+ return Exp->getType().getNonReferenceType().getCanonicalType();
+
+ // For array-sections, we need to find the type of one element of
+ // the section.
+ const auto *OASE = cast<ArraySectionExpr>(Exp);
+
+ QualType BaseType = ArraySectionExpr::getBaseOriginalType(OASE->getBase());
+
+ QualType ElemTy;
+ if (const auto *ATy = BaseType->getAsArrayTypeUnsafe())
+ ElemTy = ATy->getElementType();
+ else
+ ElemTy = BaseType->getPointeeType();
+
+ ElemTy = ElemTy.getNonReferenceType().getCanonicalType();
+ return ElemTy;
+}
+
+std::pair<const Expr *, std::optional<size_t>>
+OMPClauseMappableExprCommon::findAttachPtrExpr(
+ MappableExprComponentListRef Components, OpenMPDirectiveKind CurDirKind) {
+
+ // If we only have a single component, we have a map like "map(p)", which
+ // cannot have a base-pointer.
+ if (Components.size() < 2)
+ return {nullptr, std::nullopt};
+
+ // Only check for non-contiguous sections on target_update, since we can
+ // assume array-sections are contiguous on maps on other constructs, even if
+ // we are not sure of it at compile-time, like for a[1:x][2].
+ if (Components.back().isNonContiguous() && CurDirKind == OMPD_target_update)
+ return {nullptr, std::nullopt};
+
+ // To find the attach base-pointer, we start with the second component,
+ // stripping away one component at a time, until we reach a pointer Expr
+ // (that is not a binary operator). The first such pointer should be the
+ // attach base-pointer for the component list.
+ for (size_t I = 1; I < Components.size(); ++I) {
+ const Expr *CurExpr = Components[I].getAssociatedExpression();
+ if (!CurExpr)
+ break;
+
+ // If CurExpr is something like `p + 10`, we need to ignore it, since
+ // we are looking for `p`.
+ if (isa<BinaryOperator>(CurExpr))
+ continue;
+
+ // Keep going until we reach an Expr of pointer type.
+ QualType CurType = getComponentExprElementType(CurExpr);
+ if (!CurType->isPointerType())
+ continue;
+
+ // We have found a pointer Expr. This must be the attach pointer.
+ return {CurExpr, Components.size() - I};
+ }
+
+ return {nullptr, std::nullopt};
+}
+
OMPMapClause *OMPMapClause::Create(
const ASTContext &C, const OMPVarListLocTy &Locs, ArrayRef<Expr *> Vars,
ArrayRef<ValueDecl *> Declarations,
diff --git a/clang/lib/Basic/OpenMPKinds.cpp b/clang/lib/Basic/OpenMPKinds.cpp
index 220b31b0f19bc..2f2a5b66e4ca5 100644
--- a/clang/lib/Basic/OpenMPKinds.cpp
+++ b/clang/lib/Basic/OpenMPKinds.cpp
@@ -650,6 +650,11 @@ bool clang::isOpenMPTargetDataManagementDirective(OpenMPDirectiveKind DKind) {
DKind == OMPD_target_exit_data || DKind == OMPD_target_update;
}
+bool clang::isOpenMPTargetMapEnteringDirective(OpenMPDirectiveKind DKind) {
+ return DKind == OMPD_target_data || DKind == OMPD_target_enter_data ||
+ isOpenMPTargetExecutionDirective(DKind);
+}
+
bool clang::isOpenMPNestingTeamsDirective(OpenMPDirectiveKind DKind) {
if (DKind == OMPD_teams)
return true;
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
index f98339d472fa9..d592c29a412a9 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -6765,12 +6765,256 @@ llvm::Value *CGOpenMPRuntime::emitNumThreadsForTargetDirective(
namespace {
LLVM_ENABLE_BITMASK_ENUMS_IN_NAMESPACE();
+/// Utility to compare expression locations.
+/// Returns true if expr-loc of LHS is less-than that of RHS.
+/// This function asserts that both expressions have valid expr-locations.
+static bool compareExprLocs(const Expr *LHS, const Expr *RHS) {
+ // Assert that neither LHS nor RHS can be null
+ assert(LHS && "LHS expression cannot be null");
+ assert(RHS && "RHS expression cannot be null");
+
+ // Get source locations
+ SourceLocation LocLHS = LHS->getExprLoc();
+ SourceLocation LocRHS = RHS->getExprLoc();
+
+ // Assert that we have valid source locations
+ assert(LocLHS.isValid() && "LHS expression must have valid source location");
+ assert(LocRHS.isValid() && "RHS expression must have valid source location");
+
+ // Compare source locations for deterministic ordering
+ bool result = LocLHS < LocRHS;
+ return result;
+}
+
// Utility to handle information from clauses associated with a given
// construct that use mappable expressions (e.g. 'map' clause, 'to' clause).
// It provides a convenient interface to obtain the information and generate
// code for that information.
class MappableExprsHandler {
public:
+ /// Custom comparator for attach-pointer expressions that compares them by
+ /// complexity (i.e. their component-depth) first, then by their expr-locs if
+ /// they are semantically different.
+ struct AttachPtrExprComparator {
+ const MappableExprsHandler *Handler;
+ // Cache of previous equality comparison results.
+ mutable llvm::DenseMap<std::pair<const Expr *, const Expr *>, bool>
+ CachedEqualityComparisons;
+
+ AttachPtrExprComparator(const MappableExprsHandler *H) : Handler(H) {}
+
+ // Return true iff LHS is "less than" RHS.
+ bool operator()(const Expr *LHS, const Expr *RHS) const {
+ if (LHS == RHS)
+ return false;
+
+ // First, compare by complexity (depth)
+ auto ItLHS = Handler->AttachPtrComponentDepthMap.find(LHS);
+ auto ItRHS = Handler->AttachPtrComponentDepthMap.find(RHS);
+
+ std::optional<size_t> DepthLHS =
+ (ItLHS != Handler->AttachPtrComponentDepthMap.end()) ? ItLHS->second
+ : std::nullopt;
+ std::optional<size_t> DepthRHS =
+ (ItRHS != Handler->AttachPtrComponentDepthMap.end()) ? ItRHS->second
+ : std::nullopt;
+
+ // std::nullopt (no attach pointer) has lowest complexity
+ if (!DepthLHS.has_value() && !DepthRHS.has_value()) {
+ // Both have same complexity, now check semantic equality
+ if (areEqual(LHS, RHS))
+ return false;
+ // Different semantically, compare by location
+ return compareExprLocs(LHS, RHS);
+ }
+ if (!DepthLHS.has_value())
+ return true; // LHS has lower complexity
+ if (!DepthRHS.has_value())
+ return false; // RHS has lower complexity
+
+ // Both have values, compare by depth (lower depth = lower complexity)
+ if (DepthLHS.value() != DepthRHS.value())
+ return DepthLHS.value() < DepthRHS.value();
+
+ // Same complexity, now check semantic equality
+ if (areEqual(LHS, RHS))
+ return false;
+ // Different semantically, compare by location
+ return compareExprLocs(LHS, RHS);
+ }
+
+ public:
+ /// Return true if \p LHS and \p RHS are semantically equal. Uses pre-cached
+ /// results, if available, otherwise does a recursive semantic comparison.
+ bool areEqual(const Expr *LHS, const Expr *RHS) const {
+ // Check cache first for faster lookup
+ auto CachedResultIt = CachedEqualityComparisons.find({LHS, RHS});
+ if (CachedResultIt != CachedEqualityComparisons.end())
+ return CachedResultIt->second;
+
+ bool ComparisonResult = areSemanticallyEqual(LHS, RHS);
+
+ // Cache the result for future lookups (both orders since semantic
+ // equality is commutative)
+ CachedEqualityComparisons[{LHS, RHS}] = ComparisonResult;
+ CachedEqualityComparisons[{RHS, LHS}] = ComparisonResult;
+ return ComparisonResult;
+ }
+
+ private:
+ /// Helper function to compare attach-pointer expressions semantically.
+ /// This function handles various expression types that can be part of an
+ /// attach-pointer.
+ /// TODO: Not urgent, but we should ideally return true when comparing
+ /// `p[10]`, `*(p + 10)`, `*(p + 5 + 5)`, `p[10:1]` etc.
+ bool areSemanticallyEqual(const Expr *LHS, const Expr *RHS) const {
+ if (LHS == RHS)
+ return true;
+
+ // If only one is null, they aren't equal
+ if (!LHS || !RHS)
+ return false;
+
+ ASTContext &Ctx = Handler->CGF.getContext();
+ // Strip away parentheses and no-op casts to get to the core expression
+ LHS = LHS->IgnoreParenNoopCasts(Ctx);
+ RHS = RHS->IgnoreParenNoopCasts(Ctx);
+
+ // Direct pointer comparison of the underlying expressions
+ if (LHS == RHS)
+ return true;
+
+ // Check if the expression classes match
+ if (LHS->getStmtClass() != RHS->getStmtClass())
+ return false;
+
+ // Handle DeclRefExpr (variable references)
+ if (const auto *LD = dyn_cast<DeclRefExpr>(LHS)) {
+ const auto *RD = dyn_cast<DeclRefExpr>(RHS);
+ if (!RD)
+ return false;
+ return LD->getDecl()->getCanonicalDecl() ==
+ RD->getDecl()->getCanonicalDecl();
+ }
+
+ // Handle ArraySubscriptExpr (array indexing like a[i])
+ if (const auto *LA = dyn_cast<ArraySubscriptExpr>(LHS)) {
+ const auto *RA = dyn_cast<ArraySubscriptExpr>(RHS);
+ if (!RA)
+ return false;
+ return areSemanticallyEqual(LA->getBase(), RA->getBase()) &&
+ areSemanticallyEqual(LA->getIdx(), RA->getIdx());
+ }
+
+ // Handle MemberExpr (member access like s.m or p->m)
+ if (const auto *LM = dyn_cast<MemberExpr>(LHS)) {
+ const auto *RM = dyn_cast<MemberExpr>(RHS);
+ if (!RM)
+ return false;
+ if (LM->getMemberDecl()->getCanonicalDecl() !=
+ RM->getMemberDecl()->getCanonicalDecl())
+ return false;
+ return areSemanticallyEqual(LM->getBase(), RM->getBase());
+ }
+
+ // Handle UnaryOperator (unary operations like *p, &x, etc.)
+ if (const auto *LU = dyn_cast<UnaryOperator>(LHS)) {
+ const auto *RU = dyn_cast<UnaryOperator>(RHS);
+ if (!RU)
+ return false;
+ if (LU->getOpcode() != RU->getOpcode())
+ return false;
+ return areSemanticallyEqual(LU->getSubExpr(), RU->getSubExpr());
+ }
+
+ // Handle BinaryOperator (binary operations like p + offset)
+ if (const auto *LB = dyn_cast<BinaryOperator>(LHS)) {
+ const auto *RB = dyn_cast<BinaryOperator>(RHS);
+ if (!RB)
+ return false;
+ if (LB->getOpcode() != RB->getOpcode())
+ return false;
+ return areSemanticallyEqual(LB->getLHS(), RB->getLHS()) &&
+ areSemanticallyEqual(LB->getRHS(), RB->getRHS());
+ }
+
+ // Handle ArraySectionExpr (array sections like a[0:1])
+ // Attach pointers should not contain array-sections, but currently we
+ // don't emit an error.
+ if (const auto *LAS = dyn_cast<ArraySectionExpr>(LHS)) {
+ const auto *RAS = dyn_cast<ArraySectionExpr>(RHS);
+ if (!RAS)
+ return false;
+ return areSemanticallyEqual(LAS->getBase(), RAS->getBase()) &&
+ areSemanticallyEqual(LAS->getLowerBound(),
+ RAS->getLowerBound()) &&
+ areSemanticallyEqual(LAS->getLength(), RAS->getLength());
+ }
+
+ // Handle CastExpr (explicit casts)
+ if (const auto *LC = dyn_cast<CastExpr>(LHS)) {
+ const auto *RC = dyn_cast<CastExpr>(RHS);
+ if (!RC)
+ return false;
+ if (LC->getCastKind() != RC->getCastKind())
+ return false;
+ return areSemanticallyEqual(LC->getSubExpr(), RC->getSubExpr());
+ }
+
+ // Handle CXXThisExpr (this pointer)
+ if (isa<CXXThisExpr>(LHS) && isa<CXXThisExpr>(RHS))
+ return true;
+
+ // Handle IntegerLiteral (integer constants)
+ if (const auto *LI = dyn_cast<IntegerLiteral>(LHS)) {
+ const auto *RI = dyn_cast<IntegerLiteral>(RHS);
+ if (!RI)
+ return false;
+ return LI->getValue() == RI->getValue();
+ }
+
+ // Handle CharacterLiteral (character constants)
+ if (const auto *LC = dyn_cast<CharacterLiteral>(LHS)) {
+ const auto *RC = dyn_cast<CharacterLiteral>(RHS);
+ if (!RC)
+ return false;
+ return LC->getValue() == RC->getValue();
+ }
+
+ // Handle FloatingLiteral (floating point constants)
+ if (const auto *LF = dyn_cast<FloatingLiteral>(LHS)) {
+ const auto *RF = dyn_cast<FloatingLiteral>(RHS);
+ if (!RF)
+ return false;
+ // Use bitwise comparison for floating point literals
+ return LF->getValue().bitwiseIsEqual(RF->getValue());
+ }
+
+ // Handle StringLiteral (string constants)
+ if (const auto *LS = dyn_cast<StringLiteral>(LHS)) {
+ const auto *RS = dyn_cast<StringLiteral>(RHS);
+ if (!RS)
+ return false;
+ return LS->getString() == RS->getString();
+ }
+
+ // Handle CXXNullPtrLiteralExpr (nullptr)
+ if (isa<CXXNullPtrLiteralExpr>(LHS) && isa<CXXNullPtrLiteralExpr>(RHS))
+ return true;
+
+ // Handle CXXBoolLiteralExpr (true/false)
+ if (const auto *LB = dyn_cast<CXXBoolLiteralExpr>(LHS)) {
+ const auto *RB = dyn_cast<CXXBoolLiteralExpr>(RHS);
+ if (!RB)
+ return false;
+ return LB->getValue() == RB->getValue();
+ }
+
+ // Fallback for other forms - use the existing comparison method
+ return Expr::isSameComparisonOperand(LHS, RHS);
+ }
+ };
+
/// Get the offset of the OMP_MAP_MEMBER_OF field.
static unsigned getFlagMemberOffset() {
unsigned Offset = 0;
@@ -6846,8 +7090,42 @@ class MappableExprsHandler {
Address LB = Address::invalid();
bool IsArraySection = false;
bool HasCompleteRecord = false;
+ // ATTACH information for delaye...
[truncated]
|
|
@llvm/pr-subscribers-clang Author: Abhinav Gaba (abhinavgaba) ChangesThese have been pulled out of the codegen PR #153683, to reduce the size of that PR. Patch is 25.43 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/155625.diff 5 Files Affected:
diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h
index 1118d3e062e68..9627e99a306b4 100644
--- a/clang/include/clang/AST/OpenMPClause.h
+++ b/clang/include/clang/AST/OpenMPClause.h
@@ -5815,6 +5815,12 @@ class OMPClauseMappableExprCommon {
ValueDecl *getAssociatedDeclaration() const {
return AssociatedDeclaration;
}
+
+ bool operator==(const MappableComponent &Other) const {
+ return AssociatedExpressionNonContiguousPr ==
+ Other.AssociatedExpressionNonContiguousPr &&
+ AssociatedDeclaration == Other.AssociatedDeclaration;
+ }
};
// List of components of an expression. This first one is the whole
@@ -5828,6 +5834,95 @@ class OMPClauseMappableExprCommon {
using MappableExprComponentLists = SmallVector<MappableExprComponentList, 8>;
using MappableExprComponentListsRef = ArrayRef<MappableExprComponentList>;
+ // Hash function to allow usage as DenseMap keys.
+ friend llvm::hash_code hash_value(const MappableComponent &MC) {
+ return llvm::hash_combine(MC.getAssociatedExpression(),
+ MC.getAssociatedDeclaration(),
+ MC.isNonContiguous());
+ }
+
+public:
+ /// Get the type of an element of a ComponentList Expr \p Exp.
+ ///
+ /// For something like the following:
+ /// ```c
+ /// int *p, **p;
+ /// ```
+ /// The types for the following Exprs would be:
+ /// Expr | Type
+ /// ---------|-----------
+ /// p | int *
+ /// *p | int
+ /// p[0] | int
+ /// p[0:1] | int
+ /// pp | int **
+ /// pp[0] | int *
+ /// pp[0:1] | int *
+ /// Note: this assumes that if \p Exp is an array-section, it is contiguous.
+ static QualType getComponentExprElementType(const Expr *Exp);
+
+ /// Find the attach pointer expression from a list of mappable expression
+ /// components.
+ ///
+ /// This function traverses the component list to find the first
+ /// expression that has a pointer type, which represents the attach
+ /// base pointer expr for the current component-list.
+ ///
+ /// For example, given the following:
+ ///
+ /// ```c
+ /// struct S {
+ /// int a;
+ /// int b[10];
+ /// int c[10][10];
+ /// int *p;
+ /// int **pp;
+ /// }
+ /// S s, *ps, **pps, *(pas[10]), ***ppps;
+ /// int i;
+ /// ```
+ ///
+ /// The base-pointers for the following map operands would be:
+ /// map list-item | attach base-pointer | attach base-pointer
+ /// | for directives except | target_update (if
+ /// | target_update | different)
+ /// ----------------|-----------------------|---------------------
+ /// s | N/A |
+ /// s.a | N/A |
+ /// s.p | N/A |
+ /// ps | N/A |
+ /// ps->p | ps |
+ /// ps[1] | ps |
+ /// *(ps + 1) | ps |
+ /// (ps + 1)[1] | ps |
+ /// ps[1:10] | ps |
+ /// ps->b[10] | ps |
+ /// ps->p[10] | ps->p |
+ /// ps->c[1][2] | ps |
+ /// ps->c[1:2][2] | (error diagnostic) | N/A, TODO: ps
+ /// ps->c[1:1][2] | ps | N/A, TODO: ps
+ /// pps[1][2] | pps[1] |
+ /// pps[1:1][2] | pps[1:1] | N/A, TODO: pps[1:1]
+ /// pps[1:i][2] | pps[1:i] | N/A, TODO: pps[1:i]
+ /// pps[1:2][2] | (error diagnostic) | N/A
+ /// pps[1]->p | pps[1] |
+ /// pps[1]->p[10] | pps[1] |
+ /// pas[1] | N/A |
+ /// pas[1][2] | pas[1] |
+ /// ppps[1][2] | ppps[1] |
+ /// ppps[1][2][3] | ppps[1][2] |
+ /// ppps[1][2:1][3] | ppps[1][2:1] | N/A, TODO: ppps[1][2:1]
+ /// ppps[1][2:2][3] | (error diagnostic) | N/A
+ /// Returns a pair of the attach pointer expression and its depth in the
+ /// component list.
+ /// TODO: This may need to be updated to handle ref_ptr/ptee cases for byref
+ /// map operands.
+ /// TODO: Handle cases for target-update, where the list-item is a
+ /// non-contiguous array-section that still has a base-pointer.
+ static std::pair<const Expr *, std::optional<size_t>>
+ findAttachPtrExpr(MappableExprComponentListRef Components,
+ OpenMPDirectiveKind CurDirKind);
+
protected:
// Return the total number of elements in a list of component lists.
static unsigned
diff --git a/clang/include/clang/Basic/OpenMPKinds.h b/clang/include/clang/Basic/OpenMPKinds.h
index f40db4c13c55a..e37887e8b86ba 100644
--- a/clang/include/clang/Basic/OpenMPKinds.h
+++ b/clang/include/clang/Basic/OpenMPKinds.h
@@ -301,6 +301,14 @@ bool isOpenMPTargetExecutionDirective(OpenMPDirectiveKind DKind);
/// otherwise - false.
bool isOpenMPTargetDataManagementDirective(OpenMPDirectiveKind DKind);
+/// Checks if the specified directive is a map-entering target directive.
+/// \param DKind Specified directive.
+/// \return true - the directive is a map-entering target directive like
+/// 'omp target', 'omp target data', 'omp target enter data',
+/// 'omp target parallel', etc. (excludes 'omp target exit data', 'omp target
+/// update') otherwise - false.
+bool isOpenMPTargetMapEnteringDirective(OpenMPDirectiveKind DKind);
+
/// Checks if the specified composite/combined directive constitutes a teams
/// directive in the outermost nest. For example
/// 'omp teams distribute' or 'omp teams distribute parallel for'.
diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp
index 588b0dcc6d7b8..eff897a1a33b2 100644
--- a/clang/lib/AST/OpenMPClause.cpp
+++ b/clang/lib/AST/OpenMPClause.cpp
@@ -15,6 +15,7 @@
#include "clang/AST/Attr.h"
#include "clang/AST/Decl.h"
#include "clang/AST/DeclOpenMP.h"
+#include "clang/AST/ExprOpenMP.h"
#include "clang/Basic/LLVM.h"
#include "clang/Basic/OpenMPKinds.h"
#include "clang/Basic/TargetInfo.h"
@@ -1156,6 +1157,73 @@ unsigned OMPClauseMappableExprCommon::getUniqueDeclarationsTotalNumber(
return UniqueDecls.size();
}
+QualType
+OMPClauseMappableExprCommon::getComponentExprElementType(const Expr *Exp) {
+ assert(!isa<OMPArrayShapingExpr>(Exp) &&
+ "Cannot get element-type from array-shaping expr.");
+
+ // Unless we are handling array-section expressions, including
+ // array-subscripts, derefs, we can rely on getType.
+ if (!isa<ArraySectionExpr>(Exp))
+ return Exp->getType().getNonReferenceType().getCanonicalType();
+
+ // For array-sections, we need to find the type of one element of
+ // the section.
+ const auto *OASE = cast<ArraySectionExpr>(Exp);
+
+ QualType BaseType = ArraySectionExpr::getBaseOriginalType(OASE->getBase());
+
+ QualType ElemTy;
+ if (const auto *ATy = BaseType->getAsArrayTypeUnsafe())
+ ElemTy = ATy->getElementType();
+ else
+ ElemTy = BaseType->getPointeeType();
+
+ ElemTy = ElemTy.getNonReferenceType().getCanonicalType();
+ return ElemTy;
+}
+
+std::pair<const Expr *, std::optional<size_t>>
+OMPClauseMappableExprCommon::findAttachPtrExpr(
+ MappableExprComponentListRef Components, OpenMPDirectiveKind CurDirKind) {
+
+ // If we only have a single component, we have a map like "map(p)", which
+ // cannot have a base-pointer.
+ if (Components.size() < 2)
+ return {nullptr, std::nullopt};
+
+ // Only check for non-contiguous sections on target_update, since we can
+ // assume array-sections are contiguous on maps on other constructs, even if
+ // we are not sure of it at compile-time, like for a[1:x][2].
+ if (Components.back().isNonContiguous() && CurDirKind == OMPD_target_update)
+ return {nullptr, std::nullopt};
+
+ // To find the attach base-pointer, we start with the second component,
+ // stripping away one component at a time, until we reach a pointer Expr
+ // (that is not a binary operator). The first such pointer should be the
+ // attach base-pointer for the component list.
+ for (size_t I = 1; I < Components.size(); ++I) {
+ const Expr *CurExpr = Components[I].getAssociatedExpression();
+ if (!CurExpr)
+ break;
+
+ // If CurExpr is something like `p + 10`, we need to ignore it, since
+ // we are looking for `p`.
+ if (isa<BinaryOperator>(CurExpr))
+ continue;
+
+ // Keep going until we reach an Expr of pointer type.
+ QualType CurType = getComponentExprElementType(CurExpr);
+ if (!CurType->isPointerType())
+ continue;
+
+ // We have found a pointer Expr. This must be the attach pointer.
+ return {CurExpr, Components.size() - I};
+ }
+
+ return {nullptr, std::nullopt};
+}
+
OMPMapClause *OMPMapClause::Create(
const ASTContext &C, const OMPVarListLocTy &Locs, ArrayRef<Expr *> Vars,
ArrayRef<ValueDecl *> Declarations,
diff --git a/clang/lib/Basic/OpenMPKinds.cpp b/clang/lib/Basic/OpenMPKinds.cpp
index 220b31b0f19bc..2f2a5b66e4ca5 100644
--- a/clang/lib/Basic/OpenMPKinds.cpp
+++ b/clang/lib/Basic/OpenMPKinds.cpp
@@ -650,6 +650,11 @@ bool clang::isOpenMPTargetDataManagementDirective(OpenMPDirectiveKind DKind) {
DKind == OMPD_target_exit_data || DKind == OMPD_target_update;
}
+bool clang::isOpenMPTargetMapEnteringDirective(OpenMPDirectiveKind DKind) {
+ return DKind == OMPD_target_data || DKind == OMPD_target_enter_data ||
+ isOpenMPTargetExecutionDirective(DKind);
+}
+
bool clang::isOpenMPNestingTeamsDirective(OpenMPDirectiveKind DKind) {
if (DKind == OMPD_teams)
return true;
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
index f98339d472fa9..d592c29a412a9 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -6765,12 +6765,256 @@ llvm::Value *CGOpenMPRuntime::emitNumThreadsForTargetDirective(
namespace {
LLVM_ENABLE_BITMASK_ENUMS_IN_NAMESPACE();
+/// Utility to compare expression locations.
+/// Returns true if expr-loc of LHS is less-than that of RHS.
+/// This function asserts that both expressions have valid expr-locations.
+static bool compareExprLocs(const Expr *LHS, const Expr *RHS) {
+ // Assert that neither LHS nor RHS can be null
+ assert(LHS && "LHS expression cannot be null");
+ assert(RHS && "RHS expression cannot be null");
+
+ // Get source locations
+ SourceLocation LocLHS = LHS->getExprLoc();
+ SourceLocation LocRHS = RHS->getExprLoc();
+
+ // Assert that we have valid source locations
+ assert(LocLHS.isValid() && "LHS expression must have valid source location");
+ assert(LocRHS.isValid() && "RHS expression must have valid source location");
+
+ // Compare source locations for deterministic ordering
+ bool result = LocLHS < LocRHS;
+ return result;
+}
+
// Utility to handle information from clauses associated with a given
// construct that use mappable expressions (e.g. 'map' clause, 'to' clause).
// It provides a convenient interface to obtain the information and generate
// code for that information.
class MappableExprsHandler {
public:
+ /// Custom comparator for attach-pointer expressions that compares them by
+ /// complexity (i.e. their component-depth) first, then by their expr-locs if
+ /// they are semantically different.
+ struct AttachPtrExprComparator {
+ const MappableExprsHandler *Handler;
+ // Cache of previous equality comparison results.
+ mutable llvm::DenseMap<std::pair<const Expr *, const Expr *>, bool>
+ CachedEqualityComparisons;
+
+ AttachPtrExprComparator(const MappableExprsHandler *H) : Handler(H) {}
+
+ // Return true iff LHS is "less than" RHS.
+ bool operator()(const Expr *LHS, const Expr *RHS) const {
+ if (LHS == RHS)
+ return false;
+
+ // First, compare by complexity (depth)
+ auto ItLHS = Handler->AttachPtrComponentDepthMap.find(LHS);
+ auto ItRHS = Handler->AttachPtrComponentDepthMap.find(RHS);
+
+ std::optional<size_t> DepthLHS =
+ (ItLHS != Handler->AttachPtrComponentDepthMap.end()) ? ItLHS->second
+ : std::nullopt;
+ std::optional<size_t> DepthRHS =
+ (ItRHS != Handler->AttachPtrComponentDepthMap.end()) ? ItRHS->second
+ : std::nullopt;
+
+ // std::nullopt (no attach pointer) has lowest complexity
+ if (!DepthLHS.has_value() && !DepthRHS.has_value()) {
+ // Both have same complexity, now check semantic equality
+ if (areEqual(LHS, RHS))
+ return false;
+ // Different semantically, compare by location
+ return compareExprLocs(LHS, RHS);
+ }
+ if (!DepthLHS.has_value())
+ return true; // LHS has lower complexity
+ if (!DepthRHS.has_value())
+ return false; // RHS has lower complexity
+
+ // Both have values, compare by depth (lower depth = lower complexity)
+ if (DepthLHS.value() != DepthRHS.value())
+ return DepthLHS.value() < DepthRHS.value();
+
+ // Same complexity, now check semantic equality
+ if (areEqual(LHS, RHS))
+ return false;
+ // Different semantically, compare by location
+ return compareExprLocs(LHS, RHS);
+ }
+
+ public:
+ /// Return true if \p LHS and \p RHS are semantically equal. Uses pre-cached
+ /// results, if available, otherwise does a recursive semantic comparison.
+ bool areEqual(const Expr *LHS, const Expr *RHS) const {
+ // Check cache first for faster lookup
+ auto CachedResultIt = CachedEqualityComparisons.find({LHS, RHS});
+ if (CachedResultIt != CachedEqualityComparisons.end())
+ return CachedResultIt->second;
+
+ bool ComparisonResult = areSemanticallyEqual(LHS, RHS);
+
+ // Cache the result for future lookups (both orders since semantic
+ // equality is commutative)
+ CachedEqualityComparisons[{LHS, RHS}] = ComparisonResult;
+ CachedEqualityComparisons[{RHS, LHS}] = ComparisonResult;
+ return ComparisonResult;
+ }
+
+ private:
+ /// Helper function to compare attach-pointer expressions semantically.
+ /// This function handles various expression types that can be part of an
+ /// attach-pointer.
+ /// TODO: Not urgent, but we should ideally return true when comparing
+ /// `p[10]`, `*(p + 10)`, `*(p + 5 + 5)`, `p[10:1]` etc.
+ bool areSemanticallyEqual(const Expr *LHS, const Expr *RHS) const {
+ if (LHS == RHS)
+ return true;
+
+ // If only one is null, they aren't equal
+ if (!LHS || !RHS)
+ return false;
+
+ ASTContext &Ctx = Handler->CGF.getContext();
+ // Strip away parentheses and no-op casts to get to the core expression
+ LHS = LHS->IgnoreParenNoopCasts(Ctx);
+ RHS = RHS->IgnoreParenNoopCasts(Ctx);
+
+ // Direct pointer comparison of the underlying expressions
+ if (LHS == RHS)
+ return true;
+
+ // Check if the expression classes match
+ if (LHS->getStmtClass() != RHS->getStmtClass())
+ return false;
+
+ // Handle DeclRefExpr (variable references)
+ if (const auto *LD = dyn_cast<DeclRefExpr>(LHS)) {
+ const auto *RD = dyn_cast<DeclRefExpr>(RHS);
+ if (!RD)
+ return false;
+ return LD->getDecl()->getCanonicalDecl() ==
+ RD->getDecl()->getCanonicalDecl();
+ }
+
+ // Handle ArraySubscriptExpr (array indexing like a[i])
+ if (const auto *LA = dyn_cast<ArraySubscriptExpr>(LHS)) {
+ const auto *RA = dyn_cast<ArraySubscriptExpr>(RHS);
+ if (!RA)
+ return false;
+ return areSemanticallyEqual(LA->getBase(), RA->getBase()) &&
+ areSemanticallyEqual(LA->getIdx(), RA->getIdx());
+ }
+
+ // Handle MemberExpr (member access like s.m or p->m)
+ if (const auto *LM = dyn_cast<MemberExpr>(LHS)) {
+ const auto *RM = dyn_cast<MemberExpr>(RHS);
+ if (!RM)
+ return false;
+ if (LM->getMemberDecl()->getCanonicalDecl() !=
+ RM->getMemberDecl()->getCanonicalDecl())
+ return false;
+ return areSemanticallyEqual(LM->getBase(), RM->getBase());
+ }
+
+ // Handle UnaryOperator (unary operations like *p, &x, etc.)
+ if (const auto *LU = dyn_cast<UnaryOperator>(LHS)) {
+ const auto *RU = dyn_cast<UnaryOperator>(RHS);
+ if (!RU)
+ return false;
+ if (LU->getOpcode() != RU->getOpcode())
+ return false;
+ return areSemanticallyEqual(LU->getSubExpr(), RU->getSubExpr());
+ }
+
+ // Handle BinaryOperator (binary operations like p + offset)
+ if (const auto *LB = dyn_cast<BinaryOperator>(LHS)) {
+ const auto *RB = dyn_cast<BinaryOperator>(RHS);
+ if (!RB)
+ return false;
+ if (LB->getOpcode() != RB->getOpcode())
+ return false;
+ return areSemanticallyEqual(LB->getLHS(), RB->getLHS()) &&
+ areSemanticallyEqual(LB->getRHS(), RB->getRHS());
+ }
+
+ // Handle ArraySectionExpr (array sections like a[0:1])
+ // Attach pointers should not contain array-sections, but currently we
+ // don't emit an error.
+ if (const auto *LAS = dyn_cast<ArraySectionExpr>(LHS)) {
+ const auto *RAS = dyn_cast<ArraySectionExpr>(RHS);
+ if (!RAS)
+ return false;
+ return areSemanticallyEqual(LAS->getBase(), RAS->getBase()) &&
+ areSemanticallyEqual(LAS->getLowerBound(),
+ RAS->getLowerBound()) &&
+ areSemanticallyEqual(LAS->getLength(), RAS->getLength());
+ }
+
+ // Handle CastExpr (explicit casts)
+ if (const auto *LC = dyn_cast<CastExpr>(LHS)) {
+ const auto *RC = dyn_cast<CastExpr>(RHS);
+ if (!RC)
+ return false;
+ if (LC->getCastKind() != RC->getCastKind())
+ return false;
+ return areSemanticallyEqual(LC->getSubExpr(), RC->getSubExpr());
+ }
+
+ // Handle CXXThisExpr (this pointer)
+ if (isa<CXXThisExpr>(LHS) && isa<CXXThisExpr>(RHS))
+ return true;
+
+ // Handle IntegerLiteral (integer constants)
+ if (const auto *LI = dyn_cast<IntegerLiteral>(LHS)) {
+ const auto *RI = dyn_cast<IntegerLiteral>(RHS);
+ if (!RI)
+ return false;
+ return LI->getValue() == RI->getValue();
+ }
+
+ // Handle CharacterLiteral (character constants)
+ if (const auto *LC = dyn_cast<CharacterLiteral>(LHS)) {
+ const auto *RC = dyn_cast<CharacterLiteral>(RHS);
+ if (!RC)
+ return false;
+ return LC->getValue() == RC->getValue();
+ }
+
+ // Handle FloatingLiteral (floating point constants)
+ if (const auto *LF = dyn_cast<FloatingLiteral>(LHS)) {
+ const auto *RF = dyn_cast<FloatingLiteral>(RHS);
+ if (!RF)
+ return false;
+ // Use bitwise comparison for floating point literals
+ return LF->getValue().bitwiseIsEqual(RF->getValue());
+ }
+
+ // Handle StringLiteral (string constants)
+ if (const auto *LS = dyn_cast<StringLiteral>(LHS)) {
+ const auto *RS = dyn_cast<StringLiteral>(RHS);
+ if (!RS)
+ return false;
+ return LS->getString() == RS->getString();
+ }
+
+ // Handle CXXNullPtrLiteralExpr (nullptr)
+ if (isa<CXXNullPtrLiteralExpr>(LHS) && isa<CXXNullPtrLiteralExpr>(RHS))
+ return true;
+
+ // Handle CXXBoolLiteralExpr (true/false)
+ if (const auto *LB = dyn_cast<CXXBoolLiteralExpr>(LHS)) {
+ const auto *RB = dyn_cast<CXXBoolLiteralExpr>(RHS);
+ if (!RB)
+ return false;
+ return LB->getValue() == RB->getValue();
+ }
+
+ // Fallback for other forms - use the existing comparison method
+ return Expr::isSameComparisonOperand(LHS, RHS);
+ }
+ };
+
/// Get the offset of the OMP_MAP_MEMBER_OF field.
static unsigned getFlagMemberOffset() {
unsigned Offset = 0;
@@ -6846,8 +7090,42 @@ class MappableExprsHandler {
Address LB = Address::invalid();
bool IsArraySection = false;
bool HasCompleteRecord = false;
+ // ATTACH information for delaye...
[truncated]
|
abhinavgaba
commented
Aug 27, 2025
| /// attach-pointer. | ||
| /// TODO: Not urgent, but we should ideally return true when comparing | ||
| /// `p[10]`, `*(p + 10)`, `*(p + 5 + 5)`, `p[10:1]` etc. | ||
| bool areSemanticallyEqual(const Expr *LHS, const Expr *RHS) const { |
There was a problem hiding this comment.
@alexey-bataev, is there a cleaner/less expensive way to compare two exprs? I am trying to cache the results so that we don't do this for the same two exprs repeatedly, but the first one comparison would involve this recursive check.
There was a problem hiding this comment.
There's the FoldingSetNodeID/Profile functions, but I don't know if it would be an abuse of those to use them in this situation?
There was a problem hiding this comment.
Also I don't think those are quite enough to handle extending to compare p[10] vs. *(p + 10), etc. -- so an OpenMP-specific semantic equivalence method written out long-hand doesn't seem unreasonable to me (unless there's existing code elsewhere I'm not aware of).
kparzysz
reviewed
Aug 27, 2025
clang/lib/AST/OpenMPClause.cpp
Outdated
| // stripping away one component at a time, until we reach a pointer Expr | ||
| // (that is not a binary operator). The first such pointer should be the | ||
| // attach base-pointer for the component list. | ||
| for (size_t I = 1; I < Components.size(); ++I) { |
There was a problem hiding this comment.
Could use llvm::enumerate here
jtb20
reviewed
Aug 28, 2025
|
Having these functions without actual use is meaningless; better to split the patch by the supported types or other criteria, but with the actual uses |
|
There is already a patch that uses these functions. I guess an explicit PR stack could make this clearer, but then this PR would have to be recreated in the main repo. |
…undant variable, add some asserts.
That's going to be a bit complicated. Trying to break the change based on functional components was leaving clang in a bad intermediate state. That's why #145454 was pulled into the main PR #153683. Similarly, use_device_ptr/addr handling was breaking because of the previous code tightly coupled with the previous mapping scheme, so those had to be fixed as part of the same change. FWIW, the biggest portion of the changes in #145454 is from LIT test changes. I'll try to add more comments to the PR to link to various tests that are relevant to the code in each code-path etc. The standalone end-to-end tests were merged via different PRs, and most of the executable tests are just removing XFAILs, or doing minor updates to the CHECKs to remove the distinction between the expected/current output. |
|
@kparzysz, @alexey-bataev, I've marked the main PR (#153683) ready for review, with comments linking out to the tests that are related to the code change. I leave it to you to decide whether to review this NFC PR with the utilities, or directly review the other one. Note that majority of lines/files modified in the main PR are for tests, and the code changes are not that large, if we set aside the utilities from this NFC PR. |
|
@alexey-bataev, this is the NFC PR that we were talking about this morning. The functional PR that uses these (which I'll update once this is merged, to reduce the code size, is #153683). |
| assert(LocRHS.isValid() && "RHS expression must have valid source location"); | ||
|
|
||
| // Compare source locations for deterministic ordering | ||
| return LocLHS < LocRHS; |
There was a problem hiding this comment.
Will it work if they come from different files?
There was a problem hiding this comment.
I was trying to think of a case where the exprs we are comparing are from different files, but I couldn't think of any for attach-ptr exprs. Do you have a case in mind?
For the purpose for attach-ptr expr comparisons, we do the comparison on a per directive basis, as part of MappableExprHandler, and the ExprLocs are those for the expressions within the clause list of that directive. For example:
#pragma omp target data map(s1.p1[0], s1.p2[2])Here s1.p1 and s1.p2's ExprLocs should point to the occurrence of these expressions within this directive.
I know that we insert inlined maps for outer structs with inner structs that have declare-mappers, like:
struct S {
int *p;
};
#pragma omp declare mapper(default: s1) map(s1.p[0:1])
struct T {
S s1;
int *q;
};
T *tp;
#pragma omp target data map(tp[0]), map(from: tp[0].q) // implicit: map(tp[0].s1)However, in this case, the attach-pointer-expr for the implicit map for tp[0].s1 will still be tp, whose expr location should come from map(tp[0]) on the line with pragma omp target data
There was a problem hiding this comment.
@alexey-bataev, did you have any follow-up feedback on this?
Comment on lines
7206
to
7219
| mutable llvm::DenseMap< | ||
| OMPClauseMappableExprCommon::MappableExprComponentListRef, const Expr *> | ||
| AttachPtrExprMap; | ||
|
|
||
| /// Map from attach pointer expressions to their component depth. | ||
| /// nullptr key has std::nullopt depth. This can be used to order attach-ptr | ||
| /// expressions with increasing/decreasing depth. | ||
| /// The component-depth of `nullptr` (i.e. no attach-ptr) is `std::nullopt`. | ||
| /// TODO: Not urgent, but we should ideally use the number of pointer | ||
| /// dereferences in an expr as an indicator of its complexity, instead of the | ||
| /// component-depth. That would be needed for us to treat `p[10]`, | ||
| /// `*(p + 10)`, `*(p + 5 + 5)` together. | ||
| mutable llvm::DenseMap<const Expr *, std::optional<size_t>> | ||
| AttachPtrComponentDepthMap = {{nullptr, std::nullopt}}; |
There was a problem hiding this comment.
Why do they need to be mutable?
There was a problem hiding this comment.
Thanks! These two didn't need to be mutable. CachedEqualityComparisons had to be because it was updated in a const function.
abhinavgaba
commented
Sep 11, 2025
Comment on lines
7206
to
7219
| mutable llvm::DenseMap< | ||
| OMPClauseMappableExprCommon::MappableExprComponentListRef, const Expr *> | ||
| AttachPtrExprMap; | ||
|
|
||
| /// Map from attach pointer expressions to their component depth. | ||
| /// nullptr key has std::nullopt depth. This can be used to order attach-ptr | ||
| /// expressions with increasing/decreasing depth. | ||
| /// The component-depth of `nullptr` (i.e. no attach-ptr) is `std::nullopt`. | ||
| /// TODO: Not urgent, but we should ideally use the number of pointer | ||
| /// dereferences in an expr as an indicator of its complexity, instead of the | ||
| /// component-depth. That would be needed for us to treat `p[10]`, | ||
| /// `*(p + 10)`, `*(p + 5 + 5)` together. | ||
| mutable llvm::DenseMap<const Expr *, std::optional<size_t>> | ||
| AttachPtrComponentDepthMap = {{nullptr, std::nullopt}}; |
There was a problem hiding this comment.
Thanks! These two didn't need to be mutable. CachedEqualityComparisons had to be because it was updated in a const function.
| assert(LocRHS.isValid() && "RHS expression must have valid source location"); | ||
|
|
||
| // Compare source locations for deterministic ordering | ||
| return LocLHS < LocRHS; |
There was a problem hiding this comment.
I was trying to think of a case where the exprs we are comparing are from different files, but I couldn't think of any for attach-ptr exprs. Do you have a case in mind?
For the purpose for attach-ptr expr comparisons, we do the comparison on a per directive basis, as part of MappableExprHandler, and the ExprLocs are those for the expressions within the clause list of that directive. For example:
#pragma omp target data map(s1.p1[0], s1.p2[2])Here s1.p1 and s1.p2's ExprLocs should point to the occurrence of these expressions within this directive.
I know that we insert inlined maps for outer structs with inner structs that have declare-mappers, like:
struct S {
int *p;
};
#pragma omp declare mapper(default: s1) map(s1.p[0:1])
struct T {
S s1;
int *q;
};
T *tp;
#pragma omp target data map(tp[0]), map(from: tp[0].q) // implicit: map(tp[0].s1)However, in this case, the attach-pointer-expr for the implicit map for tp[0].s1 will still be tp, whose expr location should come from map(tp[0]) on the line with pragma omp target data
Comment on lines
+6795
to
+6796
| /// complexity (i.e. their component-depth) first, then by their expr-locs if | ||
| /// they are semantically different. |
There was a problem hiding this comment.
Why do we want to compare by the expr-locs at all?
There was a problem hiding this comment.
We need a deterministic way to order two Exprs that have the same complexity but are semantically unequal. Do you have an alternative suggestion?
Using pointer comparison cannot be reliable since that can lead to build to build differences as the pointer values can change.
There was a problem hiding this comment.
Why just the order, in which they are specified, does not work?
There was a problem hiding this comment.
Do you mean always returning true? That would introduce a logical inconsistency, if the comparison is made from two different places in different order:
a < b == true
b < a == true The exprLoc comparison will deterministically return the same result based on which clause was specified first by the user.
There was a problem hiding this comment.
Does it really matter? It will provide the stable logical dependency within a single construct; this is enough. Location-based ordering may cause trouble, if not now, then potentially in the future
There was a problem hiding this comment.
If we return true unconditionally for two un-equal EXPRs, it violates the C++ Compare requirements, which mandate that:
If comp(a, b) == true then comp(b, a) == false.
And that means the it cannot be used as a custom comparator for containers like maps etc.
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
These have been pulled out of the codegen PR #153683, to reduce the size of that PR.