diff --git a/src/common/device_helpers.hip.h b/src/common/device_helpers.hip.h
index 23d44fbdd..365126465 100644
--- a/src/common/device_helpers.hip.h
+++ b/src/common/device_helpers.hip.h
@@ -173,9 +173,11 @@ inline size_t MaxSharedMemory(int device_idx) {
 
 inline size_t MaxSharedMemoryOptin(int device_idx) {
   int max_shared_memory = 0;
+#if 0 /* CUDA Only */
   dh::safe_cuda(hipDeviceGetAttribute
                 (&max_shared_memory, hipDeviceAttributeSharedMemPerBlockOptin,
                  device_idx));
+#endif
   return static_cast<std::size_t>(max_shared_memory);
 }
 
diff --git a/src/tree/gpu_hist/histogram.cu b/src/tree/gpu_hist/histogram.cu
index 985b52c8f..7ecf825db 100644
--- a/src/tree/gpu_hist/histogram.cu
+++ b/src/tree/gpu_hist/histogram.cu
@@ -294,7 +294,7 @@ void BuildGradientHistogram(CUDAContext const* ctx, EllpackDeviceAccessor const&
 #if defined(XGBOOST_USE_CUDA)
       dh::safe_cuda(cudaFuncSetAttribute(kernel, cudaFuncAttributeMaxDynamicSharedMemorySize,
                                          max_shared_memory));
-#elif defined(XGBOOST_USE_HIP)
+#elif defined(XGBOOST_USE_HIP) && 0 /* CUDA Only */
       dh::safe_cuda(hipFuncSetAttribute((const void *)kernel, hipFuncAttributeMaxDynamicSharedMemorySize,
                                          max_shared_memory));
 #endif