From ad93a00a4947d62ea5a3e31e15977d09e694d6ed Mon Sep 17 00:00:00 2001
From: Georgii Evtushenko <evtushenko.georgy@gmail.com>
Date: Tue, 28 Nov 2023 15:33:23 -0800
Subject: [PATCH] Fix icc / cub (#1152)

Co-authored-by: Michael Schellenberger Costa <miscco@nvidia.com>
---
 cub/cub/util_macro.cuh                                 | 10 +++++++---
 .../cuda/std/detail/libcxx/include/__cccl_config       | 10 +++++-----
 2 files changed, 12 insertions(+), 8 deletions(-)

diff --git a/cub/cub/util_macro.cuh b/cub/cub/util_macro.cuh
index df2b3cd2ff..7956f655fd 100644
--- a/cub/cub/util_macro.cuh
+++ b/cub/cub/util_macro.cuh
@@ -32,7 +32,7 @@
 
 #pragma once
 
-#include <cub/version.cuh>
+#include <cuda/__cccl_config>
 
 #if defined(_CCCL_IMPLICIT_SYSTEM_HEADER_GCC)
 #  pragma GCC system_header
@@ -134,9 +134,13 @@ constexpr __host__ __device__ auto max CUB_PREVENT_MACRO_SUBSTITUTION(T &&t,
 #if !defined(CUB_DISABLE_KERNEL_VISIBILITY_WARNING_SUPPRESSION)
 _CCCL_DIAG_SUPPRESS_GCC("-Wattributes")
 _CCCL_DIAG_SUPPRESS_CLANG("-Wattributes")
-#if !defined(_CCCL_CUDA_COMPILER_NVHPC)
+#  if !defined(_CCCL_CUDA_COMPILER_NVHPC)
 _CCCL_DIAG_SUPPRESS_NVHPC(attribute_requires_external_linkage)
-#endif // !_CCCL_CUDA_COMPILER_NVHPC
+#  endif // !_CCCL_CUDA_COMPILER_NVHPC
+#  if defined(_CCCL_COMPILER_ICC) || defined(_CCCL_COMPILER_ICC_LLVM)
+#    pragma warning(disable : 1890) // the "__visibility__" attribute can only appear on functions and
+                                    // variables with external linkage'
+#  endif
 #endif // !CUB_DISABLE_KERNEL_VISIBILITY_WARNING_SUPPRESSION
 
 /** @} */       // end group UtilModule
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl_config b/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl_config
index dafa7be4ad..06e8aada87 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl_config
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl_config
@@ -12,7 +12,11 @@
 #define __CCCL_CONFIG
 
 // Determine the host compiler
-#if defined(__NVCOMPILER)
+#if defined(__INTEL_LLVM_COMPILER)
+#  define _CCCL_COMPILER_ICC_LLVM
+#elif defined(__INTEL_COMPILER)
+#  define _CCCL_COMPILER_ICC
+#elif defined(__NVCOMPILER)
 #  define _CCCL_COMPILER_NVHPC
 #elif defined(__clang__)
 #  define _CCCL_COMPILER_CLANG
@@ -22,10 +26,6 @@
 #  define _CCCL_COMPILER_MSVC
 #elif defined(__IBMCPP__)
 #  define _CCCL_COMPILER_IBM
-#elif defined(__INTEL_LLVM_COMPILER)
-#  define _CCCL_COMPILER_ICC_LLVM
-#elif defined(__INTEL_COMPILER)
-#  define _CCCL_COMPILER_ICC
 #elif defined(__CUDACC_RTC__)
 #  define _CCCL_COMPILER_NVRTC
 #endif