Fix(build): resolve merge conflicts and fix GPU compilation

- cusolver.h: remove duplicate unique_cuda_ptr/unique_host_ptr
  definitions (kept in cuda.h) to fix std::unique_ptr conversion error
- force_op.h/cu/hip.cu: restore isk parameter in GPU DeltaSpin
  kernels to match updated caller in onsite_proj_tools.cpp
  Remove duplicate spin_sign declaration from merge
- spin_constrain.h/init_sc.cpp: add cached basis_type_, ks_solver_,
  nbands_ member variables for governance PARAM reduction

Author: dyzheng

source/source_base/module_container/base/third_party/cusolver.h

@@ -20,24 +20,6 @@
 namespace container {
 namespace cuSolverConnector {
 
-struct CudaDeleter {
-    void operator()(void* ptr) const noexcept {
-        if (ptr) cudaFree(ptr);
-    }
-};
-
-template <typename T>
-using unique_cuda_ptr = std::unique_ptr<T, CudaDeleter>;
-
-struct HostDeleter {
-    void operator()(void* ptr) const noexcept {
-        if (ptr) free(ptr);
-    }
-};
-
-template <typename T>
-using unique_host_ptr = std::unique_ptr<T, HostDeleter>;
-
 #if CUDA_VERSION >= 11000
 // Generic API (CUDA 11.0+)
 template <typename T>

source/source_lcao/module_deltaspin/init_sc.cpp

@@ -57,6 +57,10 @@ void spinconstrain::SpinConstrain<TK>::init_sc(double sc_thr_in,
     this->set_input_parameters(sc_thr_in, nsc_in, nsc_min_in, alpha_trial_in, sccut_in, sc_drop_thr_in,
                                sc_acceleration_mode_in, sc_acceleration_rms_thr_in);
 
+    this->basis_type_ = PARAM.inp.basis_type;
+    this->ks_solver_ = PARAM.inp.ks_solver;
+    this->nbands_ = PARAM.inp.nbands;
+
     // Step 2: Get atom/orbital/lnchi counts from UnitCell for indexing
     // atomCounts: {element_type_index -> number_of_atoms_of_this_type}
     // orbitalCounts: {element_type_index -> number_of_orbitals_per_atom}

source/source_lcao/module_deltaspin/spin_constrain.h

@@ -1042,7 +1042,10 @@ class SpinConstrain
      std::vector<ModuleBase::Vector3<double>> lcao_lambda_in_sub_;
       /// Acceleration mode parameters
       std::string sc_acceleration_mode_ = "off"; ///< "off", "first_order", "subspace"
-      double sc_acceleration_rms_thr_ = -1.0;    ///< RMS threshold (uB) to activate acceleration, <0 disables
+       double sc_acceleration_rms_thr_ = -1.0;    ///< RMS threshold (uB) to activate acceleration, <0 disables
+       std::string basis_type_;              ///< Cached basis type for LCAO/PW branching
+       std::string ks_solver_;               ///< Cached KS solver name
+       int nbands_ = 0;                      ///< Cached number of bands
        bool acceleration_active_ = false;          ///< Has acceleration been activated this SCF iteration?
        bool acceleration_subspace_built_ = false;  ///< Has subspace been built at activation lambda?
        bool subspace_just_activated_ = false;      ///< Was subspace just activated? (signals BFGS to reset search direction)

source/source_pw/module_pwdft/kernels/cuda/force_op.cu

@@ -545,11 +545,16 @@ void cal_force_nl_op<FPTYPE, base_device::DEVICE_GPU>::operator()(const base_dev
                                                                   const FPTYPE& tpiba,
                                                                   const FPTYPE* d_wg,
                                                                   const FPTYPE* lambda,
+                                                                  const int* isk,
                                                                   const std::complex<FPTYPE>* becp,
                                                                   const std::complex<FPTYPE>* dbecp,
                                                                   FPTYPE* force)
 {
     int spin_sign = 1;
+    if (isk != nullptr && npol == 1)
+    {
+        spin_sign = (isk[ik] == 0) ? 1 : -1;
+    }
     if (npol == 1)
     {
         cal_force_onsite<FPTYPE, 1>

source/source_pw/module_pwdft/kernels/force_op.h

@@ -278,6 +278,7 @@ struct cal_force_nl_op<FPTYPE, base_device::DEVICE_GPU>
                     const FPTYPE& tpiba,
                     const FPTYPE* d_wg,
                     const FPTYPE* lambda,
+                    const int* isk,
                     const std::complex<FPTYPE>* becp,
                     const std::complex<FPTYPE>* dbecp,
                     FPTYPE* force);

source/source_pw/module_pwdft/kernels/rocm/force_op.hip.cu

@@ -539,11 +539,16 @@ void cal_force_nl_op<FPTYPE, base_device::DEVICE_GPU>::operator()(const base_dev
                                                                    const FPTYPE& tpiba,
                                                                    const FPTYPE* d_wg,
                                                                    const FPTYPE* lambda,
+                                                                   const int* isk,
                                                                    const std::complex<FPTYPE>* becp,
                                                                    const std::complex<FPTYPE>* dbecp,
                                                                    FPTYPE* force)
 {
     int spin_sign = 1;
+    if (isk != nullptr && npol == 1)
+    {
+        spin_sign = (isk[ik] == 0) ? 1 : -1;
+    }
     if (npol == 1)
     {
         hipLaunchKernelGGL(HIP_KERNEL_NAME(cal_force_onsite<FPTYPE, 1>),