From 714158608e00802fdff2492086398199b21919a5 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Wed, 20 Mar 2024 15:42:35 +0100
Subject: [PATCH 01/11] Update toolkit to 12.4

---
 ci/matrix.yaml | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/ci/matrix.yaml b/ci/matrix.yaml
index e651fb01227..22ff0a6e1f8 100644
--- a/ci/matrix.yaml
+++ b/ci/matrix.yaml
@@ -1,7 +1,7 @@
 
 cuda_prev_min: &cuda_prev_min '11.1'
 cuda_prev_max:  &cuda_prev_max  '11.8'
-cuda_curr: &cuda_curr '12.3'
+cuda_curr: &cuda_curr '12.4'
 
 # The GPUs to test on
 gpus:

From 77de584930cf994368033e960bc2acb706f7c6f3 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Wed, 20 Mar 2024 16:15:00 +0000
Subject: [PATCH 02/11] Also update the devcontainers

---
 .devcontainer/cuda11.1-gcc6/devcontainer.json          |  2 +-
 .devcontainer/cuda11.1-gcc7/devcontainer.json          |  2 +-
 .devcontainer/cuda11.1-gcc8/devcontainer.json          |  2 +-
 .devcontainer/cuda11.1-gcc9/devcontainer.json          |  2 +-
 .devcontainer/cuda11.1-llvm9/devcontainer.json         |  2 +-
 .devcontainer/cuda11.8-gcc11/devcontainer.json         |  2 +-
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../{cuda12.3-gcc7 => cuda12.4-gcc7}/devcontainer.json | 10 +++++-----
 .../{cuda12.3-gcc8 => cuda12.4-gcc8}/devcontainer.json | 10 +++++-----
 .../{cuda12.3-gcc9 => cuda12.4-gcc9}/devcontainer.json | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .../devcontainer.json                                  | 10 +++++-----
 .devcontainer/devcontainer.json                        | 10 +++++-----
 ci/matrix.yaml                                         |  2 +-
 23 files changed, 87 insertions(+), 87 deletions(-)
 rename .devcontainer/{cuda12.3-gcc10 => cuda12.4-gcc10}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-gcc11 => cuda12.4-gcc11}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-gcc12 => cuda12.4-gcc12}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-gcc7 => cuda12.4-gcc7}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-gcc8 => cuda12.4-gcc8}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-gcc9 => cuda12.4-gcc9}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm10 => cuda12.4-llvm10}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm11 => cuda12.4-llvm11}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm12 => cuda12.4-llvm12}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm13 => cuda12.4-llvm13}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm14 => cuda12.4-llvm14}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm15 => cuda12.4-llvm15}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm16 => cuda12.4-llvm16}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-llvm9 => cuda12.4-llvm9}/devcontainer.json (87%)
 rename .devcontainer/{cuda12.3-oneapi2023.2.0 => cuda12.4-oneapi2023.2.0}/devcontainer.json (85%)

diff --git a/.devcontainer/cuda11.1-gcc6/devcontainer.json b/.devcontainer/cuda11.1-gcc6/devcontainer.json
index d60fac101db..9faa25a6c45 100644
--- a/.devcontainer/cuda11.1-gcc6/devcontainer.json
+++ b/.devcontainer/cuda11.1-gcc6/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc6-cuda11.1-ubuntu18.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc6-cuda11.1-ubuntu18.04",
   "hostRequirements": {
     "gpu": "optional"
   },
diff --git a/.devcontainer/cuda11.1-gcc7/devcontainer.json b/.devcontainer/cuda11.1-gcc7/devcontainer.json
index 037968502db..3e9e2f70f0b 100644
--- a/.devcontainer/cuda11.1-gcc7/devcontainer.json
+++ b/.devcontainer/cuda11.1-gcc7/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc7-cuda11.1-ubuntu18.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc7-cuda11.1-ubuntu18.04",
   "hostRequirements": {
     "gpu": "optional"
   },
diff --git a/.devcontainer/cuda11.1-gcc8/devcontainer.json b/.devcontainer/cuda11.1-gcc8/devcontainer.json
index 66048367eec..3862680fc0f 100644
--- a/.devcontainer/cuda11.1-gcc8/devcontainer.json
+++ b/.devcontainer/cuda11.1-gcc8/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc8-cuda11.1-ubuntu18.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc8-cuda11.1-ubuntu18.04",
   "hostRequirements": {
     "gpu": "optional"
   },
diff --git a/.devcontainer/cuda11.1-gcc9/devcontainer.json b/.devcontainer/cuda11.1-gcc9/devcontainer.json
index 633bedc20cf..54f7b8881ed 100644
--- a/.devcontainer/cuda11.1-gcc9/devcontainer.json
+++ b/.devcontainer/cuda11.1-gcc9/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc9-cuda11.1-ubuntu18.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc9-cuda11.1-ubuntu18.04",
   "hostRequirements": {
     "gpu": "optional"
   },
diff --git a/.devcontainer/cuda11.1-llvm9/devcontainer.json b/.devcontainer/cuda11.1-llvm9/devcontainer.json
index 1edce1727d6..d875c9b6812 100644
--- a/.devcontainer/cuda11.1-llvm9/devcontainer.json
+++ b/.devcontainer/cuda11.1-llvm9/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm9-cuda11.1-ubuntu18.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm9-cuda11.1-ubuntu18.04",
   "hostRequirements": {
     "gpu": "optional"
   },
diff --git a/.devcontainer/cuda11.8-gcc11/devcontainer.json b/.devcontainer/cuda11.8-gcc11/devcontainer.json
index b58f0c6ea7b..389d6c7f099 100644
--- a/.devcontainer/cuda11.8-gcc11/devcontainer.json
+++ b/.devcontainer/cuda11.8-gcc11/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc11-cuda11.8-ubuntu22.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc11-cuda11.8-ubuntu22.04",
   "hostRequirements": {
     "gpu": "optional"
   },
diff --git a/.devcontainer/cuda12.3-gcc10/devcontainer.json b/.devcontainer/cuda12.4-gcc10/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-gcc10/devcontainer.json
rename to .devcontainer/cuda12.4-gcc10/devcontainer.json
index 83b52860d08..9968a8490d6 100644
--- a/.devcontainer/cuda12.3-gcc10/devcontainer.json
+++ b/.devcontainer/cuda12.4-gcc10/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc10-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc10-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-gcc10",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-gcc10",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "gcc",
     "CCCL_HOST_COMPILER_VERSION": "10",
-    "CCCL_BUILD_INFIX": "cuda12.3-gcc10"
+    "CCCL_BUILD_INFIX": "cuda12.4-gcc10"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-gcc10"
+  "name": "cuda12.4-gcc10"
 }
diff --git a/.devcontainer/cuda12.3-gcc11/devcontainer.json b/.devcontainer/cuda12.4-gcc11/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-gcc11/devcontainer.json
rename to .devcontainer/cuda12.4-gcc11/devcontainer.json
index bb1b5f76274..d2c26622341 100644
--- a/.devcontainer/cuda12.3-gcc11/devcontainer.json
+++ b/.devcontainer/cuda12.4-gcc11/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc11-cuda12.3-ubuntu22.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc11-cuda12.4-ubuntu22.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-gcc11",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-gcc11",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "gcc",
     "CCCL_HOST_COMPILER_VERSION": "11",
-    "CCCL_BUILD_INFIX": "cuda12.3-gcc11"
+    "CCCL_BUILD_INFIX": "cuda12.4-gcc11"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-gcc11"
+  "name": "cuda12.4-gcc11"
 }
diff --git a/.devcontainer/cuda12.3-gcc12/devcontainer.json b/.devcontainer/cuda12.4-gcc12/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-gcc12/devcontainer.json
rename to .devcontainer/cuda12.4-gcc12/devcontainer.json
index 5525d416fca..fa6e0c5e44f 100644
--- a/.devcontainer/cuda12.3-gcc12/devcontainer.json
+++ b/.devcontainer/cuda12.4-gcc12/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc12-cuda12.3-ubuntu22.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc12-cuda12.4-ubuntu22.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-gcc12",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-gcc12",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "gcc",
     "CCCL_HOST_COMPILER_VERSION": "12",
-    "CCCL_BUILD_INFIX": "cuda12.3-gcc12"
+    "CCCL_BUILD_INFIX": "cuda12.4-gcc12"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-gcc12"
+  "name": "cuda12.4-gcc12"
 }
diff --git a/.devcontainer/cuda12.3-gcc7/devcontainer.json b/.devcontainer/cuda12.4-gcc7/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-gcc7/devcontainer.json
rename to .devcontainer/cuda12.4-gcc7/devcontainer.json
index 651d0f901aa..af6fdb101ed 100644
--- a/.devcontainer/cuda12.3-gcc7/devcontainer.json
+++ b/.devcontainer/cuda12.4-gcc7/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc7-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc7-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-gcc7",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-gcc7",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "gcc",
     "CCCL_HOST_COMPILER_VERSION": "7",
-    "CCCL_BUILD_INFIX": "cuda12.3-gcc7"
+    "CCCL_BUILD_INFIX": "cuda12.4-gcc7"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-gcc7"
+  "name": "cuda12.4-gcc7"
 }
diff --git a/.devcontainer/cuda12.3-gcc8/devcontainer.json b/.devcontainer/cuda12.4-gcc8/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-gcc8/devcontainer.json
rename to .devcontainer/cuda12.4-gcc8/devcontainer.json
index 6e4b678c64b..46670d428da 100644
--- a/.devcontainer/cuda12.3-gcc8/devcontainer.json
+++ b/.devcontainer/cuda12.4-gcc8/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc8-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc8-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-gcc8",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-gcc8",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "gcc",
     "CCCL_HOST_COMPILER_VERSION": "8",
-    "CCCL_BUILD_INFIX": "cuda12.3-gcc8"
+    "CCCL_BUILD_INFIX": "cuda12.4-gcc8"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-gcc8"
+  "name": "cuda12.4-gcc8"
 }
diff --git a/.devcontainer/cuda12.3-gcc9/devcontainer.json b/.devcontainer/cuda12.4-gcc9/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-gcc9/devcontainer.json
rename to .devcontainer/cuda12.4-gcc9/devcontainer.json
index 4e895e09f1e..4005e723e0e 100644
--- a/.devcontainer/cuda12.3-gcc9/devcontainer.json
+++ b/.devcontainer/cuda12.4-gcc9/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc9-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc9-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-gcc9",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-gcc9",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "gcc",
     "CCCL_HOST_COMPILER_VERSION": "9",
-    "CCCL_BUILD_INFIX": "cuda12.3-gcc9"
+    "CCCL_BUILD_INFIX": "cuda12.4-gcc9"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-gcc9"
+  "name": "cuda12.4-gcc9"
 }
diff --git a/.devcontainer/cuda12.3-llvm10/devcontainer.json b/.devcontainer/cuda12.4-llvm10/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm10/devcontainer.json
rename to .devcontainer/cuda12.4-llvm10/devcontainer.json
index c7152827c3f..6ee5788664b 100644
--- a/.devcontainer/cuda12.3-llvm10/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm10/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm10-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm10-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm10",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm10",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "10",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm10"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm10"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm10"
+  "name": "cuda12.4-llvm10"
 }
diff --git a/.devcontainer/cuda12.3-llvm11/devcontainer.json b/.devcontainer/cuda12.4-llvm11/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm11/devcontainer.json
rename to .devcontainer/cuda12.4-llvm11/devcontainer.json
index a5ba402c262..66bd26cf4a3 100644
--- a/.devcontainer/cuda12.3-llvm11/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm11/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm11-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm11-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm11",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm11",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "11",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm11"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm11"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm11"
+  "name": "cuda12.4-llvm11"
 }
diff --git a/.devcontainer/cuda12.3-llvm12/devcontainer.json b/.devcontainer/cuda12.4-llvm12/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm12/devcontainer.json
rename to .devcontainer/cuda12.4-llvm12/devcontainer.json
index 7f880e9c41f..8889f147790 100644
--- a/.devcontainer/cuda12.3-llvm12/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm12/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm12-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm12-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm12",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm12",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "12",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm12"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm12"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm12"
+  "name": "cuda12.4-llvm12"
 }
diff --git a/.devcontainer/cuda12.3-llvm13/devcontainer.json b/.devcontainer/cuda12.4-llvm13/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm13/devcontainer.json
rename to .devcontainer/cuda12.4-llvm13/devcontainer.json
index 8377112ffb8..76faea9e205 100644
--- a/.devcontainer/cuda12.3-llvm13/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm13/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm13-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm13-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm13",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm13",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "13",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm13"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm13"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm13"
+  "name": "cuda12.4-llvm13"
 }
diff --git a/.devcontainer/cuda12.3-llvm14/devcontainer.json b/.devcontainer/cuda12.4-llvm14/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm14/devcontainer.json
rename to .devcontainer/cuda12.4-llvm14/devcontainer.json
index b2e70f206a1..58b00cc3362 100644
--- a/.devcontainer/cuda12.3-llvm14/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm14/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm14-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm14-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm14",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm14",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "14",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm14"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm14"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm14"
+  "name": "cuda12.4-llvm14"
 }
diff --git a/.devcontainer/cuda12.3-llvm15/devcontainer.json b/.devcontainer/cuda12.4-llvm15/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm15/devcontainer.json
rename to .devcontainer/cuda12.4-llvm15/devcontainer.json
index d88e3458eb7..9c926538e6d 100644
--- a/.devcontainer/cuda12.3-llvm15/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm15/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm15-cuda12.3-ubuntu22.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm15-cuda12.4-ubuntu22.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm15",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm15",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "15",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm15"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm15"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm15"
+  "name": "cuda12.4-llvm15"
 }
diff --git a/.devcontainer/cuda12.3-llvm16/devcontainer.json b/.devcontainer/cuda12.4-llvm16/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm16/devcontainer.json
rename to .devcontainer/cuda12.4-llvm16/devcontainer.json
index 896aabbcd27..9f6fcadfdbf 100644
--- a/.devcontainer/cuda12.3-llvm16/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm16/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm16-cuda12.3-ubuntu22.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm16-cuda12.4-ubuntu22.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm16",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm16",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "16",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm16"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm16"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm16"
+  "name": "cuda12.4-llvm16"
 }
diff --git a/.devcontainer/cuda12.3-llvm9/devcontainer.json b/.devcontainer/cuda12.4-llvm9/devcontainer.json
similarity index 87%
rename from .devcontainer/cuda12.3-llvm9/devcontainer.json
rename to .devcontainer/cuda12.4-llvm9/devcontainer.json
index 6a7275d1a38..d9910e20428 100644
--- a/.devcontainer/cuda12.3-llvm9/devcontainer.json
+++ b/.devcontainer/cuda12.4-llvm9/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-llvm9-cuda12.3-ubuntu20.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-llvm9-cuda12.4-ubuntu20.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-llvm9",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-llvm9",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "llvm",
     "CCCL_HOST_COMPILER_VERSION": "9",
-    "CCCL_BUILD_INFIX": "cuda12.3-llvm9"
+    "CCCL_BUILD_INFIX": "cuda12.4-llvm9"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-llvm9"
+  "name": "cuda12.4-llvm9"
 }
diff --git a/.devcontainer/cuda12.3-oneapi2023.2.0/devcontainer.json b/.devcontainer/cuda12.4-oneapi2023.2.0/devcontainer.json
similarity index 85%
rename from .devcontainer/cuda12.3-oneapi2023.2.0/devcontainer.json
rename to .devcontainer/cuda12.4-oneapi2023.2.0/devcontainer.json
index f04656008ea..04d71c2b080 100644
--- a/.devcontainer/cuda12.3-oneapi2023.2.0/devcontainer.json
+++ b/.devcontainer/cuda12.4-oneapi2023.2.0/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-oneapi2023.2.0-cuda12.3-ubuntu22.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-oneapi2023.2.0-cuda12.4-ubuntu22.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-oneapi2023.2.0",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-oneapi2023.2.0",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "oneapi",
     "CCCL_HOST_COMPILER_VERSION": "2023.2.0",
-    "CCCL_BUILD_INFIX": "cuda12.3-oneapi2023.2.0"
+    "CCCL_BUILD_INFIX": "cuda12.4-oneapi2023.2.0"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-oneapi2023.2.0"
+  "name": "cuda12.4-oneapi2023.2.0"
 }
diff --git a/.devcontainer/devcontainer.json b/.devcontainer/devcontainer.json
index 5525d416fca..fa6e0c5e44f 100644
--- a/.devcontainer/devcontainer.json
+++ b/.devcontainer/devcontainer.json
@@ -1,6 +1,6 @@
 {
   "shutdownAction": "stopContainer",
-  "image": "rapidsai/devcontainers:24.04-cpp-gcc12-cuda12.3-ubuntu22.04",
+  "image": "rapidsai/devcontainers:24.06-cpp-gcc12-cuda12.4-ubuntu22.04",
   "hostRequirements": {
     "gpu": "optional"
   },
@@ -14,11 +14,11 @@
     "SCCACHE_BUCKET": "rapids-sccache-devs",
     "VAULT_HOST": "https://vault.ops.k8s.rapids.ai",
     "HISTFILE": "${containerWorkspaceFolder}/.cache/._bash_history",
-    "DEVCONTAINER_NAME": "cuda12.3-gcc12",
-    "CCCL_CUDA_VERSION": "12.3",
+    "DEVCONTAINER_NAME": "cuda12.4-gcc12",
+    "CCCL_CUDA_VERSION": "12.4",
     "CCCL_HOST_COMPILER": "gcc",
     "CCCL_HOST_COMPILER_VERSION": "12",
-    "CCCL_BUILD_INFIX": "cuda12.3-gcc12"
+    "CCCL_BUILD_INFIX": "cuda12.4-gcc12"
   },
   "workspaceFolder": "/home/coder/${localWorkspaceFolderBasename}",
   "workspaceMount": "source=${localWorkspaceFolder},target=/home/coder/${localWorkspaceFolderBasename},type=bind,consistency=consistent",
@@ -42,5 +42,5 @@
       }
     }
   },
-  "name": "cuda12.3-gcc12"
+  "name": "cuda12.4-gcc12"
 }
diff --git a/ci/matrix.yaml b/ci/matrix.yaml
index 22ff0a6e1f8..12c01bac96a 100644
--- a/ci/matrix.yaml
+++ b/ci/matrix.yaml
@@ -9,7 +9,7 @@ gpus:
   - 'v100'
 
 # The version of the devcontainer images to use from https://hub.docker.com/r/rapidsai/devcontainers
-devcontainer_version: '24.04'
+devcontainer_version: '24.06'
 
 # gcc compiler configurations
 gcc6: &gcc6 { name: 'gcc', version: '6', exe: 'g++' }

From 825a052e7cf9406de617885053312019860f7b81 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Wed, 20 Mar 2024 19:03:51 +0100
Subject: [PATCH 03/11] Mark tests as passing with nvrtc

---
 .../indirectly_copyable.subsumption.compile.pass.cpp            | 2 +-
 .../indirectly_movable.subsumption.compile.pass.cpp             | 2 +-
 .../iterator.concept.bidir/subsumption.compile.pass.cpp         | 2 +-
 .../iterator.concept.forward/subsumption.compile.pass.cpp       | 2 +-
 .../iterator.concept.inc/subsumption.compile.pass.cpp           | 2 +-
 .../iterator.concept.input/subsumption.compile.pass.cpp         | 2 +-
 .../iterator.concept.iterator/subsumption.compile.pass.cpp      | 2 +-
 .../range.req/range.view/view.subsumption.compile.pass.cpp      | 2 +-
 8 files changed, 8 insertions(+), 8 deletions(-)

diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.copy/indirectly_copyable.subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.copy/indirectly_copyable.subsumption.compile.pass.cpp
index 79cc50c8528..cbfd288cc70 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.copy/indirectly_copyable.subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.copy/indirectly_copyable.subsumption.compile.pass.cpp
@@ -8,7 +8,7 @@
 //===----------------------------------------------------------------------===//
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // template<class In, class Out>
diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.move/indirectly_movable.subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.move/indirectly_movable.subsumption.compile.pass.cpp
index a4a0984927b..e62f34d8fdf 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.move/indirectly_movable.subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/alg.req.ind.move/indirectly_movable.subsumption.compile.pass.cpp
@@ -8,7 +8,7 @@
 //===----------------------------------------------------------------------===//
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // template<class In, class Out>
diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.bidir/subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.bidir/subsumption.compile.pass.cpp
index f1979c6f55b..89fe24cdefe 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.bidir/subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.bidir/subsumption.compile.pass.cpp
@@ -8,7 +8,7 @@
 //===----------------------------------------------------------------------===//
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // template<class T>
diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.forward/subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.forward/subsumption.compile.pass.cpp
index 10fa4a6be6c..74ca44e4d90 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.forward/subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.forward/subsumption.compile.pass.cpp
@@ -8,7 +8,7 @@
 //===----------------------------------------------------------------------===//
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // cuda::std::forward_iterator;
diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.inc/subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.inc/subsumption.compile.pass.cpp
index 6d08ca9fa7f..68630c2d9b7 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.inc/subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.inc/subsumption.compile.pass.cpp
@@ -8,7 +8,7 @@
 //===----------------------------------------------------------------------===//
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // template<class T>
diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.input/subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.input/subsumption.compile.pass.cpp
index 84a6bbd0f3e..98df933d80a 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.input/subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.input/subsumption.compile.pass.cpp
@@ -8,7 +8,7 @@
 //===----------------------------------------------------------------------===//
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // template<class T>
diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.iterator/subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.iterator/subsumption.compile.pass.cpp
index ec0f212eb6a..fe3afeadd89 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.iterator/subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.requirements/iterator.concepts/iterator.concept.iterator/subsumption.compile.pass.cpp
@@ -8,7 +8,7 @@
 //===----------------------------------------------------------------------===//
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // template<class In>
diff --git a/libcudacxx/test/libcudacxx/std/ranges/range.req/range.view/view.subsumption.compile.pass.cpp b/libcudacxx/test/libcudacxx/std/ranges/range.req/range.view/view.subsumption.compile.pass.cpp
index 2a959726dec..ac292bb7df8 100644
--- a/libcudacxx/test/libcudacxx/std/ranges/range.req/range.view/view.subsumption.compile.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/ranges/range.req/range.view/view.subsumption.compile.pass.cpp
@@ -9,7 +9,7 @@
 
 // UNSUPPORTED: c++03, c++11, c++14, c++17
 // UNSUPPORTED: msvc-19.16
-// XFAIL: nvrtc || nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
+// XFAIL: nvcc-12.0 || nvcc-12.1 || nvcc-12.2 || nvcc-12.3
 // nvbug 3885350
 
 // <ranges>

From d032c8525d1c509acd22f1d361773d8ee1122026 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Wed, 20 Mar 2024 19:23:48 +0100
Subject: [PATCH 04/11] Add the thrust include to libcu++ tests

---
 libcudacxx/test/libcudacxx/CMakeLists.txt | 56 +++++++++--------------
 1 file changed, 21 insertions(+), 35 deletions(-)

diff --git a/libcudacxx/test/libcudacxx/CMakeLists.txt b/libcudacxx/test/libcudacxx/CMakeLists.txt
index 67b596a4700..353cb4ca67f 100644
--- a/libcudacxx/test/libcudacxx/CMakeLists.txt
+++ b/libcudacxx/test/libcudacxx/CMakeLists.txt
@@ -36,6 +36,7 @@ else() # NOT LIBCUDACXX_TEST_WITH_NVRTC
   set(LIBCUDACXX_FORCE_INCLUDE "-include ${libcudacxx_SOURCE_DIR}/test/libcudacxx/force_include.h")
   set(LIBCUDACXX_CUDA_COMPILER "${CMAKE_CUDA_COMPILER}")
   set(LIBCUDACXX_CUDA_TEST_WITH_NVRTC "False")
+  set(LIBCUDACXX_TEST_COMPILER_FLAGS "-I${thrust_SOURCE_DIR}/thrust")
 endif()
 
 if (NOT MSVC AND NOT ${CMAKE_CUDA_COMPILER_ID} STREQUAL "Clang")
@@ -51,49 +52,34 @@ endif()
 
 # Intel OneAPI compiler has fast math enabled by default which breaks almost all floating point tests
 if (${CMAKE_CXX_COMPILER_ID} STREQUAL "IntelLLVM" OR ${CMAKE_CXX_COMPILER_ID} STREQUAL "Intel")
-  set(LIBCUDACXX_TEST_COMPILER_FLAGS
-    "${LIBCUDACXX_TEST_COMPILER_FLAGS} \
-    --compiler-options=-fno-fast-math")
+  string(APPEND LIBCUDACXX_TEST_COMPILER_FLAGS " --compiler-options=-fno-fast-math")
 endif()
 
 if (${CMAKE_CUDA_COMPILER_ID} STREQUAL "Clang")
   string(APPEND LIBCUDACXX_TEST_COMPILER_FLAGS
-     " -Xclang -fcuda-allow-variadic-functions"
-     " -Xclang -Wno-unused-parameter"
-     " -Wno-unknown-cuda-version")
-
-  find_package(CUDAToolkit)
+    " -Xclang -fcuda-allow-variadic-functions"
+    " -Xclang -Wno-unused-parameter"
+    " -Wno-unknown-cuda-version"
+    " ${LIBCUDACXX_FORCE_INCLUDE}"
+    " -I${libcudacxx_SOURCE_DIR}/include"
+    " ${LIBCUDACXX_WARNING_LEVEL}")
 
   string(APPEND LIBCUDACXX_TEST_LINKER_FLAGS
-    " -L${CUDAToolkit_LIBRARY_DIR} -lcuda -lcudart")
-endif()
-
-if (${CMAKE_CUDA_COMPILER_ID} STREQUAL "NVIDIA")
-  set(LIBCUDACXX_TEST_COMPILER_FLAGS
-    "${LIBCUDACXX_TEST_COMPILER_FLAGS} \
-    ${LIBCUDACXX_FORCE_INCLUDE} \
-    ${LIBCUDACXX_WARNING_LEVEL}")
-endif()
-
-if (${CMAKE_CUDA_COMPILER_ID} STREQUAL "NVHPC")
-  set(LIBCUDACXX_TEST_COMPILER_FLAGS
-    "${LIBCUDACXX_TEST_COMPILER_FLAGS} \
-    -stdpar")
-  set(LIBCUDACXX_TEST_LINKER_FLAGS
-    "${LIBCUDACXX_TEST_LINKER_FLAGS} \
-    -stdpar")
-endif()
-
-if (${CMAKE_CUDA_COMPILER_ID} STREQUAL "Clang")
-  set(LIBCUDACXX_TEST_COMPILER_FLAGS
-    "${LIBCUDACXX_TEST_COMPILER_FLAGS} \
-    ${LIBCUDACXX_FORCE_INCLUDE} \
-    -I${libcudacxx_SOURCE_DIR}/include \
-    ${LIBCUDACXX_WARNING_LEVEL}")
+    " -L${CUDAToolkit_LIBRARY_DIR}"
+    " -lcuda"
+    " -lcudart")
+elseif (${CMAKE_CUDA_COMPILER_ID} STREQUAL "NVIDIA")
+  string(APPEND LIBCUDACXX_TEST_COMPILER_FLAGS
+    " ${LIBCUDACXX_FORCE_INCLUDE}"
+    " ${LIBCUDACXX_WARNING_LEVEL}")
+elseif (${CMAKE_CUDA_COMPILER_ID} STREQUAL "NVHPC")
+  string(APPEND LIBCUDACXX_TEST_COMPILER_FLAGS
+    " -stdpar")
+  string(APPEND LIBCUDACXX_TEST_LINKER_FLAGS
+    " -stdpar")
 endif()
 
-set(LIBCUDACXX_COMPUTE_ARCHS_STRING
-  "${CMAKE_CUDA_ARCHITECTURES}")
+set(LIBCUDACXX_COMPUTE_ARCHS_STRING "${CMAKE_CUDA_ARCHITECTURES}")
 
 include(AddLLVM)
 

From 904fb3a16560e822df2a3c1687f8a0bec73b0f7d Mon Sep 17 00:00:00 2001
From: Wesley Maxey <wesley.maxey@gmail.com>
Date: Wed, 20 Mar 2024 14:11:25 -0700
Subject: [PATCH 05/11] Suppress warning 940 for unreachable code in variant

---
 libcudacxx/include/cuda/std/detail/libcxx/include/variant | 2 ++
 1 file changed, 2 insertions(+)

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/variant b/libcudacxx/include/cuda/std/detail/libcxx/include/variant
index 8774b62c35b..a3dfbe049ba 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/variant
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/variant
@@ -617,6 +617,7 @@ struct __variant
       _CUDA_VSTD::forward<_Vs>(__vs)...);
   }
 
+_CCCL_NV_DIAG_SUPPRESS(940) // Suppress no return at end of function
   // This overload is needed to tell the compiler that the recursion is indeed limited
   template <class _Visitor, class... _Vs, size_t... _ProcessedIndices, size_t... _UnprocessedIndices>
   inline _LIBCUDACXX_HIDDEN _LIBCUDACXX_INLINE_VISIBILITY static constexpr decltype(auto) __visit_impl(
@@ -640,6 +641,7 @@ struct __variant
     }
     _LIBCUDACXX_UNREACHABLE();
   }
+_CCCL_NV_DIAG_DEFAULT(940) // End suppression of no return at end of function
 
   template <class _Visitor, class... _Vs>
   inline _LIBCUDACXX_HIDDEN _LIBCUDACXX_INLINE_VISIBILITY static constexpr decltype(auto)

From d3095d4789c5a7f8da32d8f5a382821fdc7bb3dd Mon Sep 17 00:00:00 2001
From: Wesley Maxey <wesley.maxey@gmail.com>
Date: Wed, 20 Mar 2024 14:31:59 -0700
Subject: [PATCH 06/11] Use MSVC 14.39 for latest MSVC 2022 compiler in CI.

---
 ci/matrix.yaml | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/ci/matrix.yaml b/ci/matrix.yaml
index 12c01bac96a..589de44bd3c 100644
--- a/ci/matrix.yaml
+++ b/ci/matrix.yaml
@@ -37,7 +37,7 @@ llvm-newest: &llvm-newest { name: 'llvm', version: '16', exe: 'clang++' }
 # MSVC configs
 msvc2017: &msvc2017 { name: 'cl', version: '14.16', exe: 'cl++' }
 msvc2019: &msvc2019 { name: 'cl', version: '14.29', exe: 'cl++' }
-msvc2022: &msvc2022 { name: 'cl', version: '14.36', exe: 'cl++' }
+msvc2022: &msvc2022 { name: 'cl', version: '14.39', exe: 'cl++' }
 
 # oneAPI configs
 oneapi: &oneapi { name: 'oneapi', version: '2023.2.0', exe: 'icpc' }

From dc07987eb3e3f491b8548f8b7ddd0672b4291fb6 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Thu, 21 Mar 2024 08:22:30 +0100
Subject: [PATCH 07/11] Test `thrust::counting_iterator` traits in thrust not
 libcu++

---
 libcudacxx/test/libcudacxx/CMakeLists.txt      |  6 ++----
 .../iterator.traits/iterator.pass.cpp          | 17 -----------------
 thrust/testing/counting_iterator.cu            | 18 ++++++++++++++++++
 3 files changed, 20 insertions(+), 21 deletions(-)

diff --git a/libcudacxx/test/libcudacxx/CMakeLists.txt b/libcudacxx/test/libcudacxx/CMakeLists.txt
index 353cb4ca67f..e51f03973b1 100644
--- a/libcudacxx/test/libcudacxx/CMakeLists.txt
+++ b/libcudacxx/test/libcudacxx/CMakeLists.txt
@@ -36,13 +36,11 @@ else() # NOT LIBCUDACXX_TEST_WITH_NVRTC
   set(LIBCUDACXX_FORCE_INCLUDE "-include ${libcudacxx_SOURCE_DIR}/test/libcudacxx/force_include.h")
   set(LIBCUDACXX_CUDA_COMPILER "${CMAKE_CUDA_COMPILER}")
   set(LIBCUDACXX_CUDA_TEST_WITH_NVRTC "False")
-  set(LIBCUDACXX_TEST_COMPILER_FLAGS "-I${thrust_SOURCE_DIR}/thrust")
+  set(LIBCUDACXX_TEST_COMPILER_FLAGS "")
 endif()
 
 if (NOT MSVC AND NOT ${CMAKE_CUDA_COMPILER_ID} STREQUAL "Clang")
-  set(LIBCUDACXX_WARNING_LEVEL
-    "--compiler-options=-Wall \
-    --compiler-options=-Wextra")
+  set(LIBCUDACXX_WARNING_LEVEL "--compiler-options=-Wall --compiler-options=-Wextra")
 endif()
 
 # sccache cannot handle the -Fd option generationg pdb files
diff --git a/libcudacxx/test/libcudacxx/std/iterators/iterator.primitives/iterator.traits/iterator.pass.cpp b/libcudacxx/test/libcudacxx/std/iterators/iterator.primitives/iterator.traits/iterator.pass.cpp
index f5a059912e0..a9fec1cdff6 100644
--- a/libcudacxx/test/libcudacxx/std/iterators/iterator.primitives/iterator.traits/iterator.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/iterators/iterator.primitives/iterator.traits/iterator.pass.cpp
@@ -29,8 +29,6 @@
 #endif // TEST_COMPILER_MSVC
 
 #if !defined(TEST_COMPILER_NVRTC)
-#define THRUST_IGNORE_DEPRECATED_CPP_DIALECT
-#include <thrust/iterator/counting_iterator.h>
 #include <vector>
 #endif // !TEST_COMPILER_NVRTC
 
@@ -57,21 +55,6 @@ int main(int, char**)
   }
 
 #if !defined(TEST_COMPILER_NVRTC)
-  { // thrust iterator
-    typedef cuda::std::iterator_traits<thrust::counting_iterator<int>> It;
-    using category = thrust::detail::iterator_category_with_system_and_traversal<std::random_access_iterator_tag,
-                                                                                 thrust::any_system_tag,
-                                                                                 thrust::random_access_traversal_tag>;
-
-    static_assert(cuda::std::is_same<It::difference_type, ptrdiff_t>::value, "");
-    static_assert(cuda::std::is_same<It::value_type, int>::value, "");
-    static_assert(cuda::std::is_same<It::pointer, void>::value, "");
-    static_assert(cuda::std::is_same<It::reference, signed int>::value, "");
-    static_assert(cuda::std::is_same<It::iterator_category, category>::value, "");
-
-    static_assert(cuda::std::__is_cpp17_random_access_iterator<thrust::counting_iterator<int>>::value, "");
-  }
-
   { // std::vector
     typedef cuda::std::iterator_traits<typename std::vector<int>::iterator> It;
     static_assert((cuda::std::is_same<It::difference_type, std::ptrdiff_t>::value), "");
diff --git a/thrust/testing/counting_iterator.cu b/thrust/testing/counting_iterator.cu
index cb55f5358aa..c65700486bb 100644
--- a/thrust/testing/counting_iterator.cu
+++ b/thrust/testing/counting_iterator.cu
@@ -5,10 +5,28 @@
 #include <thrust/distance.h>
 #include <thrust/detail/cstdint.h>
 
+#include <cuda/std/iterator>
 #include <cuda/std/type_traits>
 
 THRUST_DISABLE_MSVC_POSSIBLE_LOSS_OF_DATA_WARNING_BEGIN
 
+// ensure that we properly support thrust::counting_iterator from cuda::std
+void test_iterator_traits()
+{
+  typedef cuda::std::iterator_traits<thrust::counting_iterator<int>> It;
+  using category = thrust::detail::iterator_category_with_system_and_traversal<std::random_access_iterator_tag,
+                                                                               thrust::any_system_tag,
+                                                                               thrust::random_access_traversal_tag>;
+
+  static_assert(cuda::std::is_same<It::difference_type, ptrdiff_t>::value, "");
+  static_assert(cuda::std::is_same<It::value_type, int>::value, "");
+  static_assert(cuda::std::is_same<It::pointer, void>::value, "");
+  static_assert(cuda::std::is_same<It::reference, signed int>::value, "");
+  static_assert(cuda::std::is_same<It::iterator_category, category>::value, "");
+
+  static_assert(cuda::std::__is_cpp17_random_access_iterator<thrust::counting_iterator<int>>::value, "");
+}
+
 template <typename T>
 void TestCountingDefaultConstructor(void)
 {

From 1f182175865813148b9153d7b2e0173990733c56 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Thu, 21 Mar 2024 07:46:58 +0000
Subject: [PATCH 08/11] Split variant test to not break ptxas

---
 ...all_operator_forwarding_multiple.pass.cpp} | 46 ------------
 ...t_call_operator_forwarding_single.pass.cpp | 72 ++++++++++++++++++
 ...all_operator_forwarding_multiple.pass.cpp} | 46 ------------
 ...t_call_operator_forwarding_single.pass.cpp | 75 +++++++++++++++++++
 4 files changed, 147 insertions(+), 92 deletions(-)
 rename libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/{visit_call_operator_forwarding.pass.cpp => visit_call_operator_forwarding_multiple.pass.cpp} (55%)
 create mode 100644 libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding_single.pass.cpp
 rename libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/{visit_call_operator_forwarding.pass.cpp => visit_call_operator_forwarding_multiple.pass.cpp} (55%)
 create mode 100644 libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding_single.pass.cpp

diff --git a/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding_multiple.pass.cpp
similarity index 55%
rename from libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding.pass.cpp
rename to libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding_multiple.pass.cpp
index 45dcec3608e..07310d4a66b 100644
--- a/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding_multiple.pass.cpp
@@ -41,40 +41,6 @@ void test_call_operator_forwarding() {
   using Fn = ForwardingCallObject;
   Fn obj{};
   const Fn &cobj = obj;
-  { // test call operator forwarding - no variant
-    cuda::std::visit(obj);
-    assert(Fn::check_call<>(CT_NonConst | CT_LValue));
-    cuda::std::visit(cobj);
-    assert(Fn::check_call<>(CT_Const | CT_LValue));
-    cuda::std::visit(cuda::std::move(obj));
-    assert(Fn::check_call<>(CT_NonConst | CT_RValue));
-    cuda::std::visit(cuda::std::move(cobj));
-    assert(Fn::check_call<>(CT_Const | CT_RValue));
-  }
-  { // test call operator forwarding - single variant, single arg
-    using V = cuda::std::variant<int>;
-    V v(42);
-    cuda::std::visit(obj, v);
-    assert(Fn::check_call<int &>(CT_NonConst | CT_LValue));
-    cuda::std::visit(cobj, v);
-    assert(Fn::check_call<int &>(CT_Const | CT_LValue));
-    cuda::std::visit(cuda::std::move(obj), v);
-    assert(Fn::check_call<int &>(CT_NonConst | CT_RValue));
-    cuda::std::visit(cuda::std::move(cobj), v);
-    assert(Fn::check_call<int &>(CT_Const | CT_RValue));
-  }
-  { // test call operator forwarding - single variant, multi arg
-    using V = cuda::std::variant<int, long, double>;
-    V v(42l);
-    cuda::std::visit(obj, v);
-    assert(Fn::check_call<long &>(CT_NonConst | CT_LValue));
-    cuda::std::visit(cobj, v);
-    assert(Fn::check_call<long &>(CT_Const | CT_LValue));
-    cuda::std::visit(cuda::std::move(obj), v);
-    assert(Fn::check_call<long &>(CT_NonConst | CT_RValue));
-    cuda::std::visit(cuda::std::move(cobj), v);
-    assert(Fn::check_call<long &>(CT_Const | CT_RValue));
-  }
   { // test call operator forwarding - multi variant, multi arg
     using V = cuda::std::variant<int, long, double>;
     using V2 = cuda::std::variant<int *, almost_string>;
@@ -101,18 +67,6 @@ void test_call_operator_forwarding() {
     cuda::std::visit(cuda::std::move(cobj), v1, v2, v3, v4);
     assert((Fn::check_call<long &, almost_string &, int &, double &>(CT_Const | CT_RValue)));
   }
-  {
-    using V = cuda::std::variant<int, long, double, int*, almost_string>;
-    V v1(42l), v2("hello"), v3(nullptr), v4(1.1);
-    cuda::std::visit(obj, v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_NonConst | CT_LValue)));
-    cuda::std::visit(cobj, v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_Const | CT_LValue)));
-    cuda::std::visit(cuda::std::move(obj), v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_NonConst | CT_RValue)));
-    cuda::std::visit(cuda::std::move(cobj), v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_Const | CT_RValue)));
-  }
 }
 
 int main(int, char**) {
diff --git a/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding_single.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding_single.pass.cpp
new file mode 100644
index 00000000000..3b6821637d3
--- /dev/null
+++ b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit/visit_call_operator_forwarding_single.pass.cpp
@@ -0,0 +1,72 @@
+//===----------------------------------------------------------------------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: c++03, c++11
+// UNSUPPORTED: msvc-19.16
+// UNSUPPORTED: clang-7, clang-8
+
+// <cuda/std/variant>
+// template <class Visitor, class... Variants>
+// constexpr see below visit(Visitor&& vis, Variants&&... vars);
+
+#include <cuda/std/cassert>
+// #include <cuda/std/memory>
+// #include <cuda/std/string>
+#include <cuda/std/type_traits>
+#include <cuda/std/utility>
+#include <cuda/std/variant>
+
+#include "test_macros.h"
+#include "variant_test_helpers.h"
+
+__host__ __device__
+void test_call_operator_forwarding() {
+  using Fn = ForwardingCallObject;
+  Fn obj{};
+  const Fn &cobj = obj;
+  { // test call operator forwarding - no variant
+    cuda::std::visit(obj);
+    assert(Fn::check_call<>(CT_NonConst | CT_LValue));
+    cuda::std::visit(cobj);
+    assert(Fn::check_call<>(CT_Const | CT_LValue));
+    cuda::std::visit(cuda::std::move(obj));
+    assert(Fn::check_call<>(CT_NonConst | CT_RValue));
+    cuda::std::visit(cuda::std::move(cobj));
+    assert(Fn::check_call<>(CT_Const | CT_RValue));
+  }
+  { // test call operator forwarding - single variant, single arg
+    using V = cuda::std::variant<int>;
+    V v(42);
+    cuda::std::visit(obj, v);
+    assert(Fn::check_call<int &>(CT_NonConst | CT_LValue));
+    cuda::std::visit(cobj, v);
+    assert(Fn::check_call<int &>(CT_Const | CT_LValue));
+    cuda::std::visit(cuda::std::move(obj), v);
+    assert(Fn::check_call<int &>(CT_NonConst | CT_RValue));
+    cuda::std::visit(cuda::std::move(cobj), v);
+    assert(Fn::check_call<int &>(CT_Const | CT_RValue));
+  }
+  { // test call operator forwarding - single variant, multi arg
+    using V = cuda::std::variant<int, long, double>;
+    V v(42l);
+    cuda::std::visit(obj, v);
+    assert(Fn::check_call<long &>(CT_NonConst | CT_LValue));
+    cuda::std::visit(cobj, v);
+    assert(Fn::check_call<long &>(CT_Const | CT_LValue));
+    cuda::std::visit(cuda::std::move(obj), v);
+    assert(Fn::check_call<long &>(CT_NonConst | CT_RValue));
+    cuda::std::visit(cuda::std::move(cobj), v);
+    assert(Fn::check_call<long &>(CT_Const | CT_RValue));
+  }
+}
+
+int main(int, char**) {
+  test_call_operator_forwarding();
+
+  return 0;
+}
diff --git a/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding_multiple.pass.cpp
similarity index 55%
rename from libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding.pass.cpp
rename to libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding_multiple.pass.cpp
index 5037098c1a5..9bae001fd24 100644
--- a/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding_multiple.pass.cpp
@@ -43,40 +43,6 @@ void test_call_operator_forwarding() {
   using Fn = ForwardingCallObject;
   Fn obj{};
   const Fn &cobj = obj;
-  { // test call operator forwarding - no variant
-    cuda::std::visit<ReturnType>(obj);
-    assert(Fn::check_call<>(CT_NonConst | CT_LValue));
-    cuda::std::visit<ReturnType>(cobj);
-    assert(Fn::check_call<>(CT_Const | CT_LValue));
-    cuda::std::visit<ReturnType>(cuda::std::move(obj));
-    assert(Fn::check_call<>(CT_NonConst | CT_RValue));
-    cuda::std::visit<ReturnType>(cuda::std::move(cobj));
-    assert(Fn::check_call<>(CT_Const | CT_RValue));
-  }
-  { // test call operator forwarding - single variant, single arg
-    using V = cuda::std::variant<int>;
-    V v(42);
-    cuda::std::visit<ReturnType>(obj, v);
-    assert(Fn::check_call<int &>(CT_NonConst | CT_LValue));
-    cuda::std::visit<ReturnType>(cobj, v);
-    assert(Fn::check_call<int &>(CT_Const | CT_LValue));
-    cuda::std::visit<ReturnType>(cuda::std::move(obj), v);
-    assert(Fn::check_call<int &>(CT_NonConst | CT_RValue));
-    cuda::std::visit<ReturnType>(cuda::std::move(cobj), v);
-    assert(Fn::check_call<int &>(CT_Const | CT_RValue));
-  }
-  { // test call operator forwarding - single variant, multi arg
-    using V = cuda::std::variant<int, long, double>;
-    V v(42l);
-    cuda::std::visit<ReturnType>(obj, v);
-    assert(Fn::check_call<long &>(CT_NonConst | CT_LValue));
-    cuda::std::visit<ReturnType>(cobj, v);
-    assert(Fn::check_call<long &>(CT_Const | CT_LValue));
-    cuda::std::visit<ReturnType>(cuda::std::move(obj), v);
-    assert(Fn::check_call<long &>(CT_NonConst | CT_RValue));
-    cuda::std::visit<ReturnType>(cuda::std::move(cobj), v);
-    assert(Fn::check_call<long &>(CT_Const | CT_RValue));
-  }
   { // test call operator forwarding - multi variant, multi arg
     using V = cuda::std::variant<int, long, double>;
     using V2 = cuda::std::variant<int *, almost_string>;
@@ -103,18 +69,6 @@ void test_call_operator_forwarding() {
     cuda::std::visit<ReturnType>(cuda::std::move(cobj), v1, v2, v3, v4);
     assert((Fn::check_call<long &, almost_string &, int &, double &>(CT_Const | CT_RValue)));
   }
-  {
-    using V = cuda::std::variant<int, long, double, int*, almost_string>;
-    V v1(42l), v2("hello"), v3(nullptr), v4(1.1);
-    cuda::std::visit<ReturnType>(obj, v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_NonConst | CT_LValue)));
-    cuda::std::visit<ReturnType>(cobj, v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_Const | CT_LValue)));
-    cuda::std::visit<ReturnType>(cuda::std::move(obj), v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_NonConst | CT_RValue)));
-    cuda::std::visit<ReturnType>(cuda::std::move(cobj), v1, v2, v3, v4);
-    assert((Fn::check_call<long &, almost_string &, int *&, double &>(CT_Const | CT_RValue)));
-  }
 }
 
 int main(int, char**) {
diff --git a/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding_single.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding_single.pass.cpp
new file mode 100644
index 00000000000..c1a567b32b6
--- /dev/null
+++ b/libcudacxx/test/libcudacxx/std/utilities/variant/variant.visit_return/visit_call_operator_forwarding_single.pass.cpp
@@ -0,0 +1,75 @@
+//===----------------------------------------------------------------------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: c++03, c++11
+// UNSUPPORTED: msvc-19.16
+// UNSUPPORTED: clang-7, clang-8
+// UNSUPPORTED: gcc-6
+
+// <cuda/std/variant>
+// template <class R, class Visitor, class... Variants>
+// constexpr R visit(Visitor&& vis, Variants&&... vars);
+
+#include <cuda/std/cassert>
+// #include <cuda/std/memory>
+// #include <cuda/std/string>
+#include <cuda/std/type_traits>
+#include <cuda/std/utility>
+#include <cuda/std/variant>
+
+#include "test_macros.h"
+#include "variant_test_helpers.h"
+
+template <typename ReturnType>
+__host__ __device__
+void test_call_operator_forwarding() {
+  using Fn = ForwardingCallObject;
+  Fn obj{};
+  const Fn &cobj = obj;
+  { // test call operator forwarding - no variant
+    cuda::std::visit<ReturnType>(obj);
+    assert(Fn::check_call<>(CT_NonConst | CT_LValue));
+    cuda::std::visit<ReturnType>(cobj);
+    assert(Fn::check_call<>(CT_Const | CT_LValue));
+    cuda::std::visit<ReturnType>(cuda::std::move(obj));
+    assert(Fn::check_call<>(CT_NonConst | CT_RValue));
+    cuda::std::visit<ReturnType>(cuda::std::move(cobj));
+    assert(Fn::check_call<>(CT_Const | CT_RValue));
+  }
+  { // test call operator forwarding - single variant, single arg
+    using V = cuda::std::variant<int>;
+    V v(42);
+    cuda::std::visit<ReturnType>(obj, v);
+    assert(Fn::check_call<int &>(CT_NonConst | CT_LValue));
+    cuda::std::visit<ReturnType>(cobj, v);
+    assert(Fn::check_call<int &>(CT_Const | CT_LValue));
+    cuda::std::visit<ReturnType>(cuda::std::move(obj), v);
+    assert(Fn::check_call<int &>(CT_NonConst | CT_RValue));
+    cuda::std::visit<ReturnType>(cuda::std::move(cobj), v);
+    assert(Fn::check_call<int &>(CT_Const | CT_RValue));
+  }
+  { // test call operator forwarding - single variant, multi arg
+    using V = cuda::std::variant<int, long, double>;
+    V v(42l);
+    cuda::std::visit<ReturnType>(obj, v);
+    assert(Fn::check_call<long &>(CT_NonConst | CT_LValue));
+    cuda::std::visit<ReturnType>(cobj, v);
+    assert(Fn::check_call<long &>(CT_Const | CT_LValue));
+    cuda::std::visit<ReturnType>(cuda::std::move(obj), v);
+    assert(Fn::check_call<long &>(CT_NonConst | CT_RValue));
+    cuda::std::visit<ReturnType>(cuda::std::move(cobj), v);
+    assert(Fn::check_call<long &>(CT_Const | CT_RValue));
+  }
+}
+
+int main(int, char**) {
+  test_call_operator_forwarding<void>();
+  test_call_operator_forwarding<int>();
+
+  return 0;
+}

From 075d763a954680f2fb46ccb2c81bfdf775b35a58 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Thu, 21 Mar 2024 16:51:52 +0100
Subject: [PATCH 09/11] Try harder to get MSVC to work

---
 libcudacxx/include/cuda/std/detail/libcxx/include/array | 5 +++++
 1 file changed, 5 insertions(+)

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/array b/libcudacxx/include/cuda/std/detail/libcxx/include/array
index 4b4ca38f1bf..cc1b1381abe 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/array
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/array
@@ -467,6 +467,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS array<_Tp, 0>
   {
     _LIBCUDACXX_ASSERT(false, "cannot call array<T, 0>::operator[] on a zero-sized array");
     _LIBCUDACXX_UNREACHABLE();
+    return *data();
   }
 
   _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const_reference
@@ -474,6 +475,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS array<_Tp, 0>
   {
     _LIBCUDACXX_ASSERT(false, "cannot call array<T, 0>::operator[] on a zero-sized array");
     _LIBCUDACXX_UNREACHABLE();
+    return *data();
   }
 
   _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 reference at(size_type)
@@ -500,12 +502,14 @@ struct _LIBCUDACXX_TEMPLATE_VIS array<_Tp, 0>
   {
     _LIBCUDACXX_ASSERT(false, "cannot call array<T, 0>::front() on a zero-sized array");
     _LIBCUDACXX_UNREACHABLE();
+    return *data();
   }
 
   _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 reference back() noexcept
   {
     _LIBCUDACXX_ASSERT(false, "cannot call array<T, 0>::back() on a zero-sized array");
     _LIBCUDACXX_UNREACHABLE();
+    return *data();
   }
 
   _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const_reference
@@ -513,6 +517,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS array<_Tp, 0>
   {
     _LIBCUDACXX_ASSERT(false, "cannot call array<T, 0>::back() on a zero-sized array");
     _LIBCUDACXX_UNREACHABLE();
+    return *data();
   }
 };
 

From 211665e5ff6b72b2295fc4b802bb4f7c4b059a1c Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Fri, 22 Mar 2024 10:35:34 +0100
Subject: [PATCH 10/11] One more try to fix MSVC empty array

---
 libcudacxx/include/cuda/std/detail/libcxx/include/array | 4 ++++
 1 file changed, 4 insertions(+)

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/array b/libcudacxx/include/cuda/std/detail/libcxx/include/array
index cc1b1381abe..dc48ccd3d11 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/array
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/array
@@ -461,6 +461,8 @@ struct _LIBCUDACXX_TEMPLATE_VIS array<_Tp, 0>
     return true;
   }
 
+_CCCL_DIAG_PUSH
+_CCCL_DIAG_SUPPRESS_MSVC(4702) // Unreachable code
   // element access:
   _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 reference
   operator[](size_type) noexcept
@@ -495,6 +497,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS array<_Tp, 0>
   {
     _LIBCUDACXX_ASSERT(false, "cannot call array<T, 0>::front() on a zero-sized array");
     _LIBCUDACXX_UNREACHABLE();
+    return *data();
   }
 
   _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const_reference
@@ -519,6 +522,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS array<_Tp, 0>
     _LIBCUDACXX_UNREACHABLE();
     return *data();
   }
+_CCCL_DIAG_POP
 };
 
 #if _CCCL_STD_VER >= 2017

From 739c4e6d012d654954c2c994cab8484e4b432ba3 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Mon, 25 Mar 2024 12:32:01 +0100
Subject: [PATCH 11/11] Just drop `operator[]` tests on empty array

---
 .../std/containers/sequences/array/indexing.pass.cpp      | 8 ++++----
 .../containers/sequences/array/indexing_const.pass.cpp    | 8 ++++----
 2 files changed, 8 insertions(+), 8 deletions(-)

diff --git a/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing.pass.cpp b/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing.pass.cpp
index 7b4ae7f2cfe..0a01d5e8e12 100644
--- a/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing.pass.cpp
@@ -45,10 +45,10 @@ __host__ __device__ TEST_CONSTEXPR_CXX14 bool tests()
             LIBCPP_ASSERT_NOEXCEPT(c[0]);
             ASSERT_SAME_TYPE(C::reference, decltype(c[0]));
             if (c.size() > (0)) { // always false
-#if !defined(TEST_COMPILER_MSVC_2017)
+#if !defined(TEST_COMPILER_MSVC)
               C::reference r = c[0];
               unused(r);
-#endif // !TEST_COMPILER_MSVC_2017
+#endif // !TEST_COMPILER_MSVC
             }
         }
         {
@@ -58,10 +58,10 @@ __host__ __device__ TEST_CONSTEXPR_CXX14 bool tests()
             LIBCPP_ASSERT_NOEXCEPT(c[0]);
             ASSERT_SAME_TYPE(C::reference, decltype(c[0]));
             if (c.size() > (0)) { // always false
-#if !defined(TEST_COMPILER_MSVC_2017)
+#if !defined(TEST_COMPILER_MSVC)
               C::reference r = c[0];
               unused(r);
-#endif // !TEST_COMPILER_MSVC_2017
+#endif // !TEST_COMPILER_MSVC
             }
         }
     }
diff --git a/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing_const.pass.cpp b/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing_const.pass.cpp
index 16be8d7fd6e..dd901b6842a 100644
--- a/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing_const.pass.cpp
+++ b/libcudacxx/test/libcudacxx/std/containers/sequences/array/indexing_const.pass.cpp
@@ -39,10 +39,10 @@ __host__ __device__ TEST_CONSTEXPR_CXX14 bool tests()
             LIBCPP_ASSERT_NOEXCEPT(c[0]);
             ASSERT_SAME_TYPE(C::const_reference, decltype(c[0]));
             if (c.size() > (0)) { // always false
-#if !defined(TEST_COMPILER_MSVC_2017)
+#if !defined(TEST_COMPILER_MSVC)
               C::const_reference r = c[0];
               unused(r);
-#endif // !TEST_COMPILER_MSVC_2017
+#endif // !TEST_COMPILER_MSVC
             }
         }
         {
@@ -52,10 +52,10 @@ __host__ __device__ TEST_CONSTEXPR_CXX14 bool tests()
             LIBCPP_ASSERT_NOEXCEPT(c[0]);
             ASSERT_SAME_TYPE(C::const_reference, decltype(c[0]));
             if (c.size() > (0)) { // always false
-#if !defined(TEST_COMPILER_MSVC_2017)
+#if !defined(TEST_COMPILER_MSVC)
               C::const_reference r = c[0];
               unused(r);
-#endif // !TEST_COMPILER_MSVC_2017
+#endif // !TEST_COMPILER_MSVC
             }
         }
     }