diff --git a/src/objective/lambdarank_obj.cc b/src/objective/lambdarank_obj.cc
index d0ff5bda5..93c694ec1 100644
--- a/src/objective/lambdarank_obj.cc
+++ b/src/objective/lambdarank_obj.cc
@@ -414,7 +414,7 @@ class LambdaRankNDCG : public LambdaRankObj<LambdaRankNDCG, ltr::NDCGCache> {
 };
 
 namespace cuda_impl {
-#if !defined(XGBOOST_USE_CUDA)
+#if !defined(XGBOOST_USE_CUDA) && !defined(XGBOOST_USE_HIP)
 void LambdaRankGetGradientNDCG(Context const*, std::int32_t, HostDeviceVector<float> const&,
                                const MetaInfo&, std::shared_ptr<ltr::NDCGCache>,
                                linalg::VectorView<double const>,  // input bias ratio
@@ -430,7 +430,7 @@ void LambdaRankUpdatePositionBias(Context const*, linalg::VectorView<double cons
                                   linalg::Vector<double>*, std::shared_ptr<ltr::RankingCache>) {
   common::AssertGPUSupport();
 }
-#endif  // !defined(XGBOOST_USE_CUDA)
+#endif  // !defined(XGBOOST_USE_CUDA) && !defined(XGBOOST_USE_HIP)
 }  // namespace cuda_impl
 
 namespace cpu_impl {
@@ -533,7 +533,7 @@ class LambdaRankMAP : public LambdaRankObj<LambdaRankMAP, ltr::MAPCache> {
   }
 };
 
-#if !defined(XGBOOST_USE_CUDA)
+#if !defined(XGBOOST_USE_CUDA) && !defined(XGBOOST_USE_HIP)
 namespace cuda_impl {
 void MAPStat(Context const*, MetaInfo const&, common::Span<std::size_t const>,
              std::shared_ptr<ltr::MAPCache>) {
@@ -549,7 +549,7 @@ void LambdaRankGetGradientMAP(Context const*, std::int32_t, HostDeviceVector<flo
   common::AssertGPUSupport();
 }
 }  // namespace cuda_impl
-#endif  // !defined(XGBOOST_USE_CUDA)
+#endif  // !defined(XGBOOST_USE_CUDA) && !defined(XGBOOST_USE_HIP)
 
 /**
  * \brief The RankNet loss.
@@ -604,7 +604,7 @@ class LambdaRankPairwise : public LambdaRankObj<LambdaRankPairwise, ltr::Ranking
   }
 };
 
-#if !defined(XGBOOST_USE_CUDA)
+#if !defined(XGBOOST_USE_CUDA) && !defined(XGBOOST_USE_HIP)
 namespace cuda_impl {
 void LambdaRankGetGradientPairwise(Context const*, std::int32_t, HostDeviceVector<float> const&,
                                    const MetaInfo&, std::shared_ptr<ltr::RankingCache>,
@@ -615,7 +615,7 @@ void LambdaRankGetGradientPairwise(Context const*, std::int32_t, HostDeviceVecto
   common::AssertGPUSupport();
 }
 }  // namespace cuda_impl
-#endif  // !defined(XGBOOST_USE_CUDA)
+#endif  // !defined(XGBOOST_USE_CUDA) && !defined(XGBOOST_USE_HIP)
 
 XGBOOST_REGISTER_OBJECTIVE(LambdaRankNDCG, LambdaRankNDCG::Name())
     .describe("LambdaRank with NDCG loss as objective")
diff --git a/src/objective/lambdarank_obj.cu b/src/objective/lambdarank_obj.cu
index 934a2aa62..38b912f1e 100644
--- a/src/objective/lambdarank_obj.cu
+++ b/src/objective/lambdarank_obj.cu
@@ -518,9 +518,9 @@ void LambdaRankGetGradientPairwise(Context const* ctx, std::int32_t iter,
   Launch(ctx, iter, predt, info, p_cache, delta, ti_plus, tj_minus, li, lj, out_gpair);
 }
 
-struct ReduceOp : thrust::binary_function<thrust::tuple<double, double> const&, thrust::tuple<double, double>
-                  const&, thrust::tuple<double, double>> {
-    thrust::tuple<double, double> __host__ XGBOOST_DEVICE operator()(thrust::tuple<double, double> const& l, thrust::tuple<double, double> const& r) {
+struct ReduceOp {
+    template <typename Tup>
+    Tup XGBOOST_DEVICE operator()(Tup const& l, Tup const& r) const {
     return thrust::make_tuple(thrust::get<0>(l) + thrust::get<0>(r),
                               thrust::get<1>(l) + thrust::get<1>(r));
   }