Merge from 'main' to 'sycl-web' (48 commits)

  CONFLICT (content): Merge conflict in clang/test/Driver/HLSL/metal-converter.hlsl
  CONFLICT (content): Merge conflict in clang/test/Driver/dxc_dxv_path.hlsl

Author: vmaksimo

clang-tools-extra/clang-tidy/bugprone/InvalidEnumDefaultInitializationCheck.cpp

@@ -22,11 +22,10 @@ bool isCompleteAndHasNoZeroValue(const EnumDecl *D) {
   const EnumDecl *Definition = D->getDefinition();
   return Definition && Definition->isComplete() &&
          !Definition->enumerators().empty() &&
-         std::none_of(Definition->enumerator_begin(),
-                      Definition->enumerator_end(),
-                      [](const EnumConstantDecl *Value) {
-                        return Value->getInitVal().isZero();
-                      });
+         llvm::none_of(Definition->enumerators(),
+                       [](const EnumConstantDecl *Value) {
+                         return Value->getInitVal().isZero();
+                       });
 }
 
 AST_MATCHER(EnumDecl, isCompleteAndHasNoZeroValue) {

clang-tools-extra/clang-tidy/bugprone/NonZeroEnumToBoolConversionCheck.cpp

@@ -22,11 +22,10 @@ namespace {
 AST_MATCHER(EnumDecl, isCompleteAndHasNoZeroValue) {
   const EnumDecl *Definition = Node.getDefinition();
   return Definition && Node.isComplete() &&
-         std::none_of(Definition->enumerator_begin(),
-                      Definition->enumerator_end(),
-                      [](const EnumConstantDecl *Value) {
-                        return Value->getInitVal().isZero();
-                      });
+         llvm::none_of(Definition->enumerators(),
+                       [](const EnumConstantDecl *Value) {
+                         return Value->getInitVal().isZero();
+                       });
 }
 
 } // namespace

clang-tools-extra/clang-tidy/utils/RenamerClangTidyCheck.cpp

@@ -432,6 +432,10 @@ RenamerClangTidyCheck::addUsage(
   if (FixLocation.isInvalid())
     return {NamingCheckFailures.end(), false};
 
+  // Skip if in system system header
+  if (SourceMgr.isInSystemHeader(FixLocation))
+    return {NamingCheckFailures.end(), false};
+
   auto EmplaceResult = NamingCheckFailures.try_emplace(FailureId);
   NamingCheckFailure &Failure = EmplaceResult.first->second;
 
@@ -455,6 +459,9 @@ RenamerClangTidyCheck::addUsage(
 void RenamerClangTidyCheck::addUsage(const NamedDecl *Decl,
                                      SourceRange UsageRange,
                                      const SourceManager &SourceMgr) {
+  if (SourceMgr.isInSystemHeader(Decl->getLocation()))
+    return;
+
   if (hasNoName(Decl))
     return;
 

clang-tools-extra/docs/ReleaseNotes.rst

@@ -130,6 +130,10 @@ Changes in existing checks
   <clang-tidy/checks/bugprone/infinite-loop>` check by adding detection for
   variables introduced by structured bindings.
 
+- Improved :doc:`bugprone-reserved-identifier
+  <clang-tidy/checks/bugprone/reserved-identifier>` check by ignoring
+  declarations in system headers.
+
 - Improved :doc:`bugprone-signed-char-misuse
   <clang-tidy/checks/bugprone/signed-char-misuse>` check by fixing
   false positives on C23 enums with the fixed underlying type of signed char.
@@ -160,6 +164,10 @@ Changes in existing checks
   <clang-tidy/checks/portability/template-virtual-member-function>` check to
   avoid false positives on pure virtual member functions.
 
+- Improved :doc:`readability-identifier-naming
+  <clang-tidy/checks/readability/identifier-naming>` check by ignoring
+  declarations in system headers.
+
 - Improved :doc:`readability-qualified-auto
   <clang-tidy/checks/readability/qualified-auto>` check by adding the option
   `IgnoreAliasing`, that allows not looking at underlying types of type aliases.

clang/docs/LanguageExtensions.rst

@@ -2018,7 +2018,7 @@ even if there is no valid ``std::tuple_element`` specialization or suitable
 Blocks
 ======
 
-The syntax and high level language feature description is in
+The syntax and high-level language feature description is in
 :doc:`BlockLanguageSpec<BlockLanguageSpec>`. Implementation and ABI details for
 the clang implementation are in :doc:`Block-ABI-Apple<Block-ABI-Apple>`.
 
@@ -2088,7 +2088,7 @@ producing an object with the following member functions
   constexpr size_t size() const;
 
 such as ``std::string``, ``std::string_view``, ``std::vector<char>``.
-This mechanism follow the same rules as ``static_assert`` messages in
+This mechanism follows the same rules as ``static_assert`` messages in
 C++26, see ``[dcl.pre]/p12``.
 
 Query for this feature with ``__has_extension(gnu_asm_constexpr_strings)``.
@@ -2335,7 +2335,7 @@ Objective-C Autosynthesis of Properties
 
 Clang provides support for autosynthesis of declared properties.  Using this
 feature, clang provides default synthesis of those properties not declared
-@dynamic and not having user provided backing getter and setter methods.
+@dynamic and not having user-provided backing getter and setter methods.
 ``__has_feature(objc_default_synthesize_properties)`` checks for availability
 of this feature in version of clang being used.
 
@@ -2349,7 +2349,7 @@ In Objective-C, functions and methods are generally assumed to follow the
 <https://developer.apple.com/library/mac/#documentation/Cocoa/Conceptual/MemoryMgmt/Articles/mmRules.html>`_
 conventions for ownership of object arguments and
 return values. However, there are exceptions, and so Clang provides attributes
-to allow these exceptions to be documented. This are used by ARC and the
+to allow these exceptions to be documented. These are used by ARC and the
 `static analyzer <https://clang-analyzer.llvm.org>`_ Some exceptions may be
 better described using the ``objc_method_family`` attribute instead.
 
@@ -2575,7 +2575,7 @@ Such functionality is not conformant and does not guarantee to compile
 correctly in any circumstances. It can be used if:
 
 - the kernel source does not contain call expressions to (member-) function
-  pointers, or virtual functions. For example this extension can be used in
+  pointers, or virtual functions. For example, this extension can be used in
   metaprogramming algorithms to be able to specify/detect types generically.
 
 - the generated kernel binary does not contain indirect calls because they
@@ -2613,7 +2613,7 @@ functions with variadic prototypes do not get generated in binary e.g. the
 variadic prototype is used to specify a function type with any number of
 arguments in metaprogramming algorithms in C++ for OpenCL.
 
-This extensions can also be used when the kernel code is intended for targets
+This extension can also be used when the kernel code is intended for targets
 supporting the variadic arguments e.g. majority of CPU targets.
 
 **Example of Use**:
@@ -2702,7 +2702,7 @@ address space qualifiers, therefore, other type qualifiers such as
 Legacy 1.x atomics with generic address space
 ---------------------------------------------
 
-Clang allows use of atomic functions from the OpenCL 1.x standards
+Clang allows the use of atomic functions from the OpenCL 1.x standards
 with the generic address space pointer in C++ for OpenCL mode.
 
 This is a non-portable feature and might not be supported by all
@@ -2833,7 +2833,7 @@ to a possibly overlapping destination region. It takes five arguments.
 The first argument is the destination WebAssembly table, and the second
 argument is the source WebAssembly table. The third argument is the
 destination index from where the copy starts, the fourth argument is the
-source index from there the copy starts, and the fifth and last argument
+source index from where the copy starts, and the fifth and last argument
 is the number of elements to copy. It returns nothing.
 
 .. code-block:: c++
@@ -3133,7 +3133,7 @@ Query for this feature with ``__has_builtin(__builtin_get_vtable_pointer)``.
 ------------------------------------
 
 ``__builtin_call_with_static_chain`` is used to perform a static call while
-setting updating the static chain register.
+updating the static chain register.
 
 **Syntax**:
 
@@ -3245,7 +3245,7 @@ Query for this feature with ``__has_builtin(__builtin_readsteadycounter)``.
 The ``__builtin_cpu_supports`` function detects if the run-time CPU supports
 features specified in string argument. It returns a positive integer if all
 features are supported and 0 otherwise. Feature names are target specific. On
-AArch64 features are combined using ``+`` like this
+AArch64, features are combined using ``+`` like this
 ``__builtin_cpu_supports("flagm+sha3+lse+rcpc2+fcma+memtag+bti+sme2")``.
 If a feature name is not supported, Clang will issue a warning and replace
 builtin by the constant 0.
@@ -3465,7 +3465,7 @@ Query for this feature with ``__has_builtin(__builtin_convertvector)``.
 **Description**:
 
 The '``__builtin_bitreverse``' family of builtins is used to reverse
-the bitpattern of an integer value; for example ``0b10110110`` becomes
+the bitpattern of an integer value; for example, ``0b10110110`` becomes
 ``0b01101101``. These builtins can be used within constant expressions.
 
 ``__builtin_rotateleft``
@@ -3970,7 +3970,7 @@ the debugging experience.
 
 ``__builtin_allow_runtime_check`` returns true if the check at the current
 program location should be executed. It is expected to be used to implement
-``assert`` like checks which can be safely removed by optimizer.
+``assert`` like checks which can be safely removed by the optimizer.
 
 **Syntax**:
 

clang/docs/ReleaseNotes.rst

@@ -123,6 +123,8 @@ Improvements to Clang's diagnostics
   Moved the warning for a missing (though implied) attribute on a redeclaration into this group.
   Added a new warning in this group for the case where the attribute is missing/implicit on
   an override of a virtual method.
+- Fixed fix-it hint for fold expressions. Clang now correctly places the suggested right 
+  parenthesis when diagnosing malformed fold expressions. (#GH151787)
 
 Improvements to Clang's time-trace
 ----------------------------------

clang/include/clang/Basic/BuiltinsAMDGPU.def

@@ -721,6 +721,12 @@ TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk8_bf16_fp4, "V8yUiUiIUi", "nc", "gfx
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk8_f32_fp8, "V8fV2UiUiIUi", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk8_f32_bf8, "V8fV2UiUiIUi", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk8_f32_fp4, "V8fUiUiIUi", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk16_f16_fp6, "V16hV3UiUiIUi", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk16_bf16_fp6, "V16yV3UiUiIUi", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk16_f16_bf6, "V16hV3UiUiIUi", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk16_bf16_bf6, "V16yV3UiUiIUi", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk16_f32_fp6, "V16fV3UiUiIUi", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scale_pk16_f32_bf6, "V16fV3UiUiIUi", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk8_fp8_bf16, "V2UiV8yf", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk8_bf8_bf16, "V2UiV8yf", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk8_fp8_f16, "V2UiV8hf", "nc", "gfx1250-insts")
@@ -730,6 +736,12 @@ TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk8_bf8_f32, "V2UiV8ff", "nc", "gfx
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk8_fp4_f32, "UiV8ff", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk8_fp4_f16, "UiV8hf", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk8_fp4_bf16, "UiV8yf", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk16_fp6_f32, "V3UiV16ff", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk16_bf6_f32, "V3UiV16ff", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk16_fp6_f16, "V3UiV16hf", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk16_bf6_f16, "V3UiV16hf", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk16_fp6_bf16, "V3UiV16yf", "nc", "gfx1250-insts")
+TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_pk16_bf6_bf16, "V3UiV16yf", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_sr_pk8_fp8_bf16, "V2UiV8yUif", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_sr_pk8_bf8_bf16, "V2UiV8yUif", "nc", "gfx1250-insts")
 TARGET_BUILTIN(__builtin_amdgcn_cvt_scalef32_sr_pk8_fp8_f16, "V2UiV8hUif", "nc", "gfx1250-insts")
@@ -750,6 +762,10 @@ TARGET_BUILTIN(__builtin_amdgcn_permlane_down, "iiii", "nc", "gfx1250-insts,wave
 TARGET_BUILTIN(__builtin_amdgcn_permlane_xor, "iiii", "nc", "gfx1250-insts,wavefrontsize32")
 TARGET_BUILTIN(__builtin_amdgcn_permlane_idx_gen, "iii", "nc", "gfx1250-insts,wavefrontsize32")
 
+TARGET_BUILTIN(__builtin_amdgcn_perm_pk16_b4_u4, "V2UiUiUiV2Ui", "nc", "tensor-cvt-lut-insts")
+TARGET_BUILTIN(__builtin_amdgcn_perm_pk16_b6_u4, "V3UiUiULiV2Ui", "nc", "tensor-cvt-lut-insts")
+TARGET_BUILTIN(__builtin_amdgcn_perm_pk16_b8_u4, "V4UiULiULiV2Ui", "nc", "tensor-cvt-lut-insts")
+
 // GFX1250 WMMA builtins
 TARGET_BUILTIN(__builtin_amdgcn_wmma_f32_16x16x4_f32, "V8fIbV2fIbV2fIsV8fIbIb", "nc", "gfx1250-insts,wavefrontsize32")
 TARGET_BUILTIN(__builtin_amdgcn_wmma_f32_16x16x32_bf16, "V8fIbV16yIbV16yIsV8fIbIb", "nc", "gfx1250-insts,wavefrontsize32")

clang/lib/AST/ByteCode/ByteCodeEmitter.cpp

@@ -137,21 +137,21 @@ int32_t ByteCodeEmitter::getOffset(LabelTy Label) {
 template <typename T>
 static void emit(Program &P, std::vector<std::byte> &Code, const T &Val,
                  bool &Success) {
+  size_t ValPos = Code.size();
   size_t Size;
 
   if constexpr (std::is_pointer_v<T>)
-    Size = sizeof(uint32_t);
+    Size = align(sizeof(uint32_t));
   else
-    Size = sizeof(T);
+    Size = align(sizeof(T));
 
-  if (Code.size() + Size > std::numeric_limits<unsigned>::max()) {
+  if (ValPos + Size > std::numeric_limits<unsigned>::max()) {
     Success = false;
     return;
   }
 
   // Access must be aligned!
-  size_t ValPos = align(Code.size());
-  Size = align(Size);
+  assert(aligned(ValPos));
   assert(aligned(ValPos + Size));
   Code.resize(ValPos + Size);
 
@@ -168,17 +168,16 @@ static void emit(Program &P, std::vector<std::byte> &Code, const T &Val,
 template <typename T>
 static void emitSerialized(std::vector<std::byte> &Code, const T &Val,
                            bool &Success) {
-  size_t Size = Val.bytesToSerialize();
+  size_t ValPos = Code.size();
+  size_t Size = align(Val.bytesToSerialize());
 
-  if (Code.size() + Size > std::numeric_limits<unsigned>::max()) {
+  if (ValPos + Size > std::numeric_limits<unsigned>::max()) {
     Success = false;
     return;
   }
 
   // Access must be aligned!
-  assert(aligned(Code.size()));
-  size_t ValPos = Code.size();
-  Size = align(Size);
+  assert(aligned(ValPos));
   assert(aligned(ValPos + Size));
   Code.resize(ValPos + Size);
 

clang/lib/AST/ByteCode/Compiler.cpp

@@ -331,6 +331,8 @@ bool Compiler<Emitter>::VisitCastExpr(const CastExpr *CE) {
   }
 
   case CK_FloatingToIntegral: {
+    if (!CE->getType()->isIntegralOrEnumerationType())
+      return false;
     if (!this->visit(SubExpr))
       return false;
     PrimType ToT = classifyPrim(CE);
@@ -1369,10 +1371,15 @@ bool Compiler<Emitter>::VisitVectorBinOp(const BinaryOperator *E) {
   // BitAdd/BitOr/BitXor/Shl/Shr doesn't support bool type, we need perform the
   // integer promotion.
   bool NeedIntPromot = ElemT == PT_Bool && (E->isBitwiseOp() || E->isShiftOp());
-  QualType PromotTy =
-      Ctx.getASTContext().getPromotedIntegerType(Ctx.getASTContext().BoolTy);
-  PrimType PromotT = classifyPrim(PromotTy);
-  PrimType OpT = NeedIntPromot ? PromotT : ElemT;
+  QualType PromotTy;
+  PrimType PromotT = PT_Bool;
+  PrimType OpT = ElemT;
+  if (NeedIntPromot) {
+    PromotTy =
+        Ctx.getASTContext().getPromotedIntegerType(Ctx.getASTContext().BoolTy);
+    PromotT = classifyPrim(PromotTy);
+    OpT = PromotT;
+  }
 
   auto getElem = [=](unsigned Offset, PrimType ElemT, unsigned Index) {
     if (!this->emitGetLocal(PT_Ptr, Offset, E))

clang/lib/AST/ByteCode/Descriptor.cpp

@@ -21,14 +21,31 @@
 using namespace clang;
 using namespace clang::interp;
 
+template <typename T> static constexpr bool needsCtor() {
+  if constexpr (std::is_same_v<T, Integral<8, true>> ||
+                std::is_same_v<T, Integral<8, false>> ||
+                std::is_same_v<T, Integral<16, true>> ||
+                std::is_same_v<T, Integral<16, false>> ||
+                std::is_same_v<T, Integral<32, true>> ||
+                std::is_same_v<T, Integral<32, false>> ||
+                std::is_same_v<T, Integral<64, true>> ||
+                std::is_same_v<T, Integral<64, false>> ||
+                std::is_same_v<T, Boolean>)
+    return false;
+
+  return true;
+}
+
 template <typename T>
 static void ctorTy(Block *, std::byte *Ptr, bool, bool, bool, bool, bool,
                    const Descriptor *) {
+  static_assert(needsCtor<T>());
   new (Ptr) T();
 }
 
 template <typename T>
 static void dtorTy(Block *, std::byte *Ptr, const Descriptor *) {
+  static_assert(needsCtor<T>());
   reinterpret_cast<T *>(Ptr)->~T();
 }
 
@@ -45,9 +62,11 @@ static void ctorArrayTy(Block *, std::byte *Ptr, bool, bool, bool, bool, bool,
                         const Descriptor *D) {
   new (Ptr) InitMapPtr(std::nullopt);
 
-  Ptr += sizeof(InitMapPtr);
-  for (unsigned I = 0, NE = D->getNumElems(); I < NE; ++I) {
-    new (&reinterpret_cast<T *>(Ptr)[I]) T();
+  if constexpr (needsCtor<T>()) {
+    Ptr += sizeof(InitMapPtr);
+    for (unsigned I = 0, NE = D->getNumElems(); I < NE; ++I) {
+      new (&reinterpret_cast<T *>(Ptr)[I]) T();
+    }
   }
 }
 
@@ -57,9 +76,12 @@ static void dtorArrayTy(Block *, std::byte *Ptr, const Descriptor *D) {
 
   if (IMP)
     IMP = std::nullopt;
-  Ptr += sizeof(InitMapPtr);
-  for (unsigned I = 0, NE = D->getNumElems(); I < NE; ++I) {
-    reinterpret_cast<T *>(Ptr)[I].~T();
+
+  if constexpr (needsCtor<T>()) {
+    Ptr += sizeof(InitMapPtr);
+    for (unsigned I = 0, NE = D->getNumElems(); I < NE; ++I) {
+      reinterpret_cast<T *>(Ptr)[I].~T();
+    }
   }
 }
 
@@ -74,10 +96,14 @@ static void moveArrayTy(Block *, std::byte *Src, std::byte *Dst,
   }
   Src += sizeof(InitMapPtr);
   Dst += sizeof(InitMapPtr);
-  for (unsigned I = 0, NE = D->getNumElems(); I < NE; ++I) {
-    auto *SrcPtr = &reinterpret_cast<T *>(Src)[I];
-    auto *DstPtr = &reinterpret_cast<T *>(Dst)[I];
-    new (DstPtr) T(std::move(*SrcPtr));
+  if constexpr (!needsCtor<T>()) {
+    std::memcpy(Dst, Src, D->getNumElems() * D->getElemSize());
+  } else {
+    for (unsigned I = 0, NE = D->getNumElems(); I < NE; ++I) {
+      auto *SrcPtr = &reinterpret_cast<T *>(Src)[I];
+      auto *DstPtr = &reinterpret_cast<T *>(Dst)[I];
+      new (DstPtr) T(std::move(*SrcPtr));
+    }
   }
 }