[Openmp-commits] [clang] [openmp] [OpenMP] Fix nondependent inscan variables in templated functions (PR #191627)

[Robert Imschweiler via Openmp-commits]

https://github.com/ro-i created https://github.com/llvm/llvm-project/pull/191627

Fixes https://github.com/llvm/llvm-project/issues/191549.

Assisted-by: claude-4.6-opus

-------------------------------

Note: I'm currently not very familiar with this portion of the code, so while the fix is tiny, I'm not sure it's the right approach. I just wanted to check the reason for the bug and since it seemed not that troublesome to fix/review, I thought that I might still go ahead and create this patch.

>From ff8463a83568ecee8caecee6be2d1236c84694b4 Mon Sep 17 00:00:00 2001
From: Robert Imschweiler <robert.imschweiler at amd.com>
Date: Sat, 11 Apr 2026 07:19:06 -0500
Subject: [PATCH] [OpenMP] Fix nondependent inscan variables in templated
 functions

Fixes https://github.com/llvm/llvm-project/issues/191549.

Assisted-by: claude-4.6-opus
---
 clang/lib/Sema/SemaOpenMP.cpp                 |  4 +
 .../scan_inscan_template_nondependent.cpp     | 63 ++++++++++++++++
 .../for/omp_scan_inscan_template.cpp          | 73 +++++++++++++++++++
 3 files changed, 140 insertions(+)
 create mode 100644 clang/test/OpenMP/scan_inscan_template_nondependent.cpp
 create mode 100644 openmp/runtime/test/worksharing/for/omp_scan_inscan_template.cpp

diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp
index 73cbc7a428661..cdeca15872a95 100644
--- a/clang/lib/Sema/SemaOpenMP.cpp
+++ b/clang/lib/Sema/SemaOpenMP.cpp
@@ -20498,6 +20498,10 @@ static bool actOnOMPReductionKindClause(
         (DeclareReductionRef.isUnset() ||
          isa<UnresolvedLookupExpr>(DeclareReductionRef.get()))) {
       RD.push(RefExpr, DeclareReductionRef.get());
+      // Handle non-dependent inscan reduction variables in dependent contexts.
+      if (RD.RedModifier == OMPC_REDUCTION_inscan)
+        Stack->addDSA(D, RefExpr->IgnoreParens(), OMPC_reduction, nullptr,
+                      RD.RedModifier, ASE || OASE);
       continue;
     }
     if (BOK == BO_Comma && DeclareReductionRef.isUnset()) {
diff --git a/clang/test/OpenMP/scan_inscan_template_nondependent.cpp b/clang/test/OpenMP/scan_inscan_template_nondependent.cpp
new file mode 100644
index 0000000000000..b56c1b3f50a8c
--- /dev/null
+++ b/clang/test/OpenMP/scan_inscan_template_nondependent.cpp
@@ -0,0 +1,63 @@
+// RUN: %clang_cc1 -verify -fopenmp -std=c++17 %s
+// RUN: %clang_cc1 -verify -fopenmp-simd -std=c++17 %s
+// expected-no-diagnostics
+
+// Regression test for https://github.com/llvm/llvm-project/issues/191549
+// Inscan reductions with non-dependent types in function templates were
+// incorrectly rejected because the DSA stack was not populated in
+// dependent contexts.
+
+template <typename T>
+void exclusive_in_template() {
+    int a[100], b[100];
+    int sum = 0;
+    #pragma omp parallel for reduction(inscan, +: sum)
+    for (int i = 0; i < 100; i++) {
+        a[i] = sum;
+        #pragma omp scan exclusive(sum)
+        sum += b[i];
+    }
+}
+
+template <typename T>
+void inclusive_in_template() {
+    int a[100], b[100];
+    int sum = 0;
+    #pragma omp parallel for reduction(inscan, +: sum)
+    for (int i = 0; i < 100; i++) {
+        sum += b[i];
+        #pragma omp scan inclusive(sum)
+        a[i] = sum;
+    }
+}
+
+template <typename T>
+void for_simd_inscan_in_template() {
+    int a[100], b[100];
+    int sum = 0;
+    #pragma omp parallel for simd reduction(inscan, +: sum)
+    for (int i = 0; i < 100; i++) {
+        a[i] = sum;
+        #pragma omp scan exclusive(sum)
+        sum += b[i];
+    }
+}
+
+template <typename T>
+void simd_inscan_in_template() {
+    int a[100], b[100];
+    int sum = 0;
+    #pragma omp simd reduction(inscan, +: sum)
+    for (int i = 0; i < 100; i++) {
+        a[i] = sum;
+        #pragma omp scan exclusive(sum)
+        sum += b[i];
+    }
+}
+
+void instantiate() {
+    exclusive_in_template<int>();
+    inclusive_in_template<int>();
+    for_simd_inscan_in_template<int>();
+    simd_inscan_in_template<int>();
+}
diff --git a/openmp/runtime/test/worksharing/for/omp_scan_inscan_template.cpp b/openmp/runtime/test/worksharing/for/omp_scan_inscan_template.cpp
new file mode 100644
index 0000000000000..685a18967b58b
--- /dev/null
+++ b/openmp/runtime/test/worksharing/for/omp_scan_inscan_template.cpp
@@ -0,0 +1,73 @@
+// RUN: %libomp-cxx-compile-and-run | FileCheck %s
+
+// Regression test for https://github.com/llvm/llvm-project/issues/191549
+// Inscan reductions with non-dependent types inside function templates were
+// rejected by Clang because the DSA stack was not populated in dependent
+// contexts.  This test verifies runtime correctness of both exclusive and
+// inclusive scans in template functions.
+
+#include <cstdlib>
+#include <numeric>
+#include <stdio.h>
+
+#define N 100
+
+template <typename T> bool test_exclusive(T *a, const T *b) {
+  int sum = 0;
+#pragma omp parallel for reduction(inscan, + : sum)
+  for (int i = 0; i < N; i++) {
+    a[i] = sum;
+#pragma omp scan exclusive(sum)
+    sum += b[i];
+  }
+
+  for (int i = 0, prefix = 0; i < N; i++) {
+    if (a[i] != prefix)
+      return false;
+    prefix += b[i];
+  }
+  return true;
+}
+
+template <typename T> bool test_inclusive(T *a, const T *b) {
+  int sum = 0;
+#pragma omp parallel for reduction(inscan, + : sum)
+  for (int i = 0; i < N; i++) {
+    sum += b[i];
+#pragma omp scan inclusive(sum)
+    a[i] = sum;
+  }
+
+  for (int i = 0, prefix = 0; i < N; i++) {
+    prefix += b[i];
+    if (a[i] != prefix)
+      return false;
+  }
+  return true;
+}
+
+int main() {
+  bool success = true;
+
+  int a[N], b[N];
+  std::iota(b, b + N, 1);
+
+  if (test_exclusive<int>(a, b)) {
+    printf("exclusive: PASS\n");
+  } else {
+    printf("exclusive: FAIL\n");
+    success = false;
+  }
+
+  if (test_inclusive<int>(a, b)) {
+    printf("inclusive: PASS\n");
+  } else {
+    printf("inclusive: FAIL\n");
+    success = false;
+  }
+
+  return success ? EXIT_SUCCESS : EXIT_FAILURE;
+}
+
+// CHECK: exclusive: PASS
+// CHECK: inclusive: PASS