Merge branch 'main' into cir-binassign

Author: mmha

clang/docs/ReleaseNotes.rst

@@ -170,6 +170,8 @@ C23 Feature Support
   scope.
 - Fixed a bug where you could not cast a null pointer constant to type
   ``nullptr_t``. Fixes #GH133644.
+- Fixed a failed assertion with an invalid parameter to the ``#embed``
+  directive. Fixes #GH126940.
 
 C11 Feature Support
 ^^^^^^^^^^^^^^^^^^^

clang/lib/AST/ItaniumMangle.cpp

@@ -2511,7 +2511,6 @@ bool CXXNameMangler::mangleUnresolvedTypeOrSimpleId(QualType Ty,
   case Type::PackIndexing:
   case Type::TemplateTypeParm:
   case Type::UnaryTransform:
-  case Type::SubstTemplateTypeParm:
   unresolvedType:
     // Some callers want a prefix before the mangled type.
     Out << Prefix;
@@ -2524,6 +2523,16 @@ bool CXXNameMangler::mangleUnresolvedTypeOrSimpleId(QualType Ty,
     // so we return directly.
     return true;
 
+  case Type::SubstTemplateTypeParm: {
+    auto *ST = cast<SubstTemplateTypeParmType>(Ty);
+    // If this was replaced from a type alias, this is not substituted
+    // from an outer template parameter, so it's not an unresolved-type.
+    if (auto *TD = dyn_cast<TemplateDecl>(ST->getAssociatedDecl());
+        TD && TD->isTypeAlias())
+      return mangleUnresolvedTypeOrSimpleId(ST->getReplacementType(), Prefix);
+    goto unresolvedType;
+  }
+
   case Type::Typedef:
     mangleSourceNameWithAbiTags(cast<TypedefType>(Ty)->getDecl());
     break;

clang/lib/Lex/PPDirectives.cpp

@@ -3897,7 +3897,9 @@ Preprocessor::LexEmbedParameters(Token &CurTok, bool ForHasEmbed) {
           return std::nullopt;
       }
       if (!ForHasEmbed) {
-        Diag(CurTok, diag::err_pp_unknown_parameter) << 1 << Parameter;
+        Diag(ParamStartLoc, diag::err_pp_unknown_parameter) << 1 << Parameter;
+        if (CurTok.isNot(tok::eod))
+          DiscardUntilEndOfDirective(CurTok);
         return std::nullopt;
       }
     }

clang/test/CodeGenCXX/mangle.cpp

@@ -1220,3 +1220,30 @@ namespace test61 {
   // CHECK-LABEL: @_ZN6test611fINS_1XEEEvNT_1Y1aENS3_1bE
   template void f<X>(int, int);
 }
+
+namespace test62 {
+  template <class> struct integral_constant {
+    static const int value = true;
+  };
+  template <int> struct _OrImpl {};
+  template <class _Args> using _Or = _OrImpl<_Args::value>;
+  template <class _Up>
+  void f(_Or<integral_constant<_Up>>) {}
+  // CHECK-LABEL: @_ZN6test621fIiEEvNS_7_OrImplIXsr17integral_constantIT_EE5valueEEE
+  template void f<int>(_OrImpl<1>);
+} // namespace test62
+
+namespace test63 {
+  namespace {
+    template <class, class> struct integral_constant {
+      static const int value = true;
+    };
+    template <class, class> struct _And {};
+    template <int> struct _OrImpl {};
+    template <class _First> using _Or = _OrImpl<_First::value>;
+    template <class _Up>
+    void f(_And<integral_constant<int, void>, _Or<integral_constant<_Up, int>>>);
+  } // namespace
+  // CHECK-LABEL: @_ZN6test6312_GLOBAL__N_11fIiEEvNS0_4_AndINS0_17integral_constantIivEENS0_7_OrImplIXsr17integral_constantIT_iEE5valueEEEEE
+  void g() { f<int>({}); }
+} // namespace test63

clang/test/CodeGenOpenCL/builtins-amdgcn-raw-buffer-load-lds.cl

@@ -0,0 +1,12 @@
+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py
+// REQUIRES: amdgpu-registered-target
+// RUN: %clang_cc1 -triple amdgcn-unknown-unknown -target-cpu gfx90a -emit-llvm -o - %s | FileCheck %s
+
+// CHECK-LABEL: @test_amdgcn_raw_ptr_buffer_load_lds(
+// CHECK-NEXT:  entry:
+// CHECK-NEXT:    tail call void @llvm.amdgcn.raw.ptr.buffer.load.lds(ptr addrspace(8) [[RSRC:%.*]], ptr addrspace(3) [[LDS:%.*]], i32 1, i32 [[OFFSET:%.*]], i32 [[SOFFSET:%.*]], i32 2, i32 3)
+// CHECK-NEXT:    ret void
+//
+void test_amdgcn_raw_ptr_buffer_load_lds(__amdgpu_buffer_rsrc_t rsrc, __local void * lds, int offset, int soffset) {
+    __builtin_amdgcn_raw_ptr_buffer_load_lds(rsrc, lds, 1, offset, soffset, 2, 3);
+}

clang/test/CodeGenOpenCL/builtins-amdgcn-raw-buffer-load.cl

@@ -170,12 +170,3 @@ v3u32 test_amdgcn_raw_ptr_buffer_load_b96_non_const_soffset(__amdgpu_buffer_rsrc
 v4u32 test_amdgcn_raw_ptr_buffer_load_b128_non_const_soffset(__amdgpu_buffer_rsrc_t rsrc, int offset, int soffset) {
   return __builtin_amdgcn_raw_buffer_load_b128(rsrc, /*offset=*/0, soffset, /*aux=*/0);
 }
-
-// CHECK-LABEL: @test_amdgcn_raw_ptr_buffer_load_lds(
-// CHECK-NEXT:  entry:
-// CHECK-NEXT:    tail call void @llvm.amdgcn.raw.ptr.buffer.load.lds(ptr addrspace(8) [[RSRC:%.*]], ptr addrspace(3) [[LDS:%.*]], i32 1, i32 [[OFFSET:%.*]], i32 [[SOFFSET:%.*]], i32 2, i32 3)
-// CHECK-NEXT:    ret void
-//
-void test_amdgcn_raw_ptr_buffer_load_lds(__amdgpu_buffer_rsrc_t rsrc, __local void * lds, int offset, int soffset) {
-    __builtin_amdgcn_raw_ptr_buffer_load_lds(rsrc, lds, 1, offset, soffset, 2, 3);
-}

clang/test/Preprocessor/embed_parameter_unrecognized.c

@@ -1,9 +1,17 @@
 // RUN: %clang_cc1 %s -std=c23 -E -verify
-// okay-no-diagnostics
 
 #embed __FILE__ unrecognized
 // expected-error@-1 {{unknown embed preprocessor parameter 'unrecognized'}}
 #embed __FILE__ unrecognized::param
 // expected-error@-1 {{unknown embed preprocessor parameter 'unrecognized::param'}}
 #embed __FILE__ unrecognized::param(with, args)
 // expected-error@-1 {{unknown embed preprocessor parameter 'unrecognized::param'}}
+
+// The first of these two cases was causing a failed assertion due to not being
+// at the end of the directive when we finished skipping past `bla`.
+#embed __FILE__ bla(2) limit(2) // expected-error {{unknown embed preprocessor parameter 'bla'}}
+#embed __FILE__ limit(2) bla(2) // expected-error {{unknown embed preprocessor parameter 'bla'}}
+
+// We intentionally only tell you about one invalid parameter at a time so that
+// we can skip over invalid token soup.
+#embed __FILE__ bla(2) limit(2) bork // expected-error {{unknown embed preprocessor parameter 'bla'}}

clang/test/SemaOpenCL/builtins-amdgcn-raw-ptr-buffer-load-lds-target-error.cl

@@ -1,4 +1,8 @@
+// RUN: %clang_cc1 -triple amdgcn-unknown-unknown -target-cpu tahiti -S -verify -o - %s
+// RUN: %clang_cc1 -triple amdgcn-unknown-unknown -target-cpu bonaire -S -verify -o - %s
+// RUN: %clang_cc1 -triple amdgcn-unknown-unknown -target-cpu carrizo -S -verify -o - %s
 // RUN: %clang_cc1 -triple amdgcn-unknown-unknown -target-cpu gfx1100 -S -verify -o - %s
+// RUN: %clang_cc1 -triple amdgcn-unknown-unknown -target-cpu gfx1200 -S -verify -o - %s
 // REQUIRES: amdgpu-registered-target
 
 void test_amdgcn_raw_ptr_buffer_load_lds(__amdgpu_buffer_rsrc_t rsrc, __local void* lds, int offset, int soffset, int x) {

compiler-rt/test/asan/TestCases/use-after-poison-history-size-partial-granule.cpp

@@ -6,6 +6,7 @@
 // RUN: %clangxx_asan -O0 %s -o %t && env ASAN_OPTIONS=poison_history_size=1000 not %run %t    2>&1 | FileCheck %s
 
 // TODO
+// REQUIRES: linux
 // UNSUPPORTED: android
 
 #include <stdio.h>

compiler-rt/test/asan/TestCases/use-after-poison-history-size.cpp

@@ -17,6 +17,7 @@
 // RUN: %clangxx_asan -O0 %s -o %t && env ASAN_OPTIONS=poison_history_size=1    not %run %t       2>&1 | FileCheck %s --check-prefixes=CHECK-ACDE,CHECK-BDE,CHECK-E
 
 // TODO
+// REQUIRES: linux
 // UNSUPPORTED: android
 
 #include <stdio.h>