!
! Copyright (C) 2021, MaX CoE
! Distributed under the MIT License
! (license terms are at http://opensource.org/licenses/MIT).
!
!--
!
! Wrappers for selected Linear Algebra subroutines
!
#include<devxlib_defs.h>
#include<devxlib_macros.h>
!
#ifdef __DXL_HAVE_DEVICE
  !
  ! checks
  !
#if ! defined __DXL_CUBLAS && ! defined __DXL_ROCBLAS && ! defined __DXL_MKL_GPU
# error
#endif
#endif

submodule (devxlib_linalg) devxlib_linalg_dot

   implicit none

   contains

   module function dev_SDOT_gpu(N,X,INCX,Y,INCY) result(res)
      implicit none
      real(real32) :: res
      integer,  intent(in) :: N
      integer,  intent(in) :: INCX,INCY
      real(real32), intent(in) DEV_ATTR :: X(:)
      real(real32)             DEV_ATTR :: Y(:)
      !
      integer :: ierr
      !
      call acc_loc_dev_sync()
      !
      !DEV_ACC data present(X,Y)
      !DEV_ACC host_data use_device(X,Y)
#if defined __DXL_CUBLAS
      res= cublasSdot(N,X,INCX,Y,INCY)
#elif defined __DXL_ROCBLAS
      !DEV_OMPGPU target data use_device_ptr(X,Y)
      ierr= rocblas_sdot(handle,N,c_loc(X),INCX,c_loc(Y),INCY,c_loc(res))
      !DEV_OMPGPU end target data
#elif defined __DXL_MKL_GPU
      !DEV_OMPGPU target data use_device_addr(X,Y)
      !DEV_OMPGPU dispatch
      res= SDOT(N,X,INCX,Y,INCY)
      !DEV_OMPGPU end dispatch
      !DEV_OMPGPU end target data
#else
      res= SDOT(N,X,INCX,Y,INCY)
#endif
      !DEV_ACC end host_data
      !DEV_ACC end data
      !
      call acc_loc_dev_sync()
      !
   end function dev_SDOT_gpu
   !
   module function dev_DDOT_gpu(N,X,INCX,Y,INCY) result(res)
      implicit none
      real(real64) :: res
      integer,  intent(in) :: N
      integer,  intent(in) :: INCX,INCY
      real(real64), intent(in) DEV_ATTR :: X(:)
      real(real64)             DEV_ATTR :: Y(:)
      !
      integer :: ierr
      !
      call acc_loc_dev_sync()
      !
      !DEV_ACC data present(X,Y)
      !DEV_ACC host_data use_device(X,Y)
#if defined __DXL_CUBLAS
      res= cublasDdot(N,X,INCX,Y,INCY)
#elif defined __DXL_ROCBLAS
      !DEV_OMPGPU target data use_device_ptr(X,Y)
      ierr= rocblas_ddot(handle,N,c_loc(X),INCX,c_loc(Y),INCY,c_loc(res))
      !DEV_OMPGPU end target data
#elif defined __DXL_MKL_GPU
      !DEV_OMPGPU target data use_device_addr(X,Y)
      !DEV_OMPGPU dispatch
      res= DDOT(N,X,INCX,Y,INCY)
      !DEV_OMPGPU end dispatch
      !DEV_OMPGPU end target data
#else
      res= DDOT(N,X,INCX,Y,INCY)
#endif
      !DEV_ACC end host_data
      !DEV_ACC end data
      !
      call acc_loc_dev_sync()
      !
   end function dev_DDOT_gpu
   !
   module function dev_CDOTU_gpu(N,X,INCX,Y,INCY) result(res)
      implicit none
      complex(real32) :: res
      integer,  intent(in) :: N
      integer,  intent(in) :: INCX,INCY
      complex(real32), intent(in) DEV_ATTR :: X(:)
      complex(real32)             DEV_ATTR :: Y(:)
      !
      integer :: ierr
      !
      call acc_loc_dev_sync()
      !
      !DEV_ACC data present(X,Y)
      !DEV_ACC host_data use_device(X,Y)
#if defined __DXL_CUBLAS
      res= cublasCdotu(N,X,INCX,Y,INCY)
#elif defined __DXL_ROCBLAS
      !DEV_OMPGPU target data use_device_ptr(X,Y)
      ierr= rocblas_cdotu(handle,N,c_loc(X),INCX,c_loc(Y),INCY,c_loc(res))
      !DEV_OMPGPU end target data
#elif defined __DXL_MKL_GPU
      !DEV_OMPGPU target data use_device_addr(X,Y)
      !DEV_OMPGPU dispatch
      res= CDOTU(N,X,INCX,Y,INCY)
      !DEV_OMPGPU end dispatch
      !DEV_OMPGPU end target data
#else
      res= CDOTU(N,X,INCX,Y,INCY)
#endif
      !DEV_ACC end host_data
      !DEV_ACC end data
      !
      call acc_loc_dev_sync()
      !
   end function dev_CDOTU_gpu
   !
   module function dev_ZDOTU_gpu(N,X,INCX,Y,INCY) result(res)
      implicit none
      complex(real64) :: res
      integer,  intent(in) :: N
      integer,  intent(in) :: INCX,INCY
      complex(real64), intent(in) DEV_ATTR :: X(:)
      complex(real64)             DEV_ATTR :: Y(:)
      !
      integer :: ierr
      !
      call acc_loc_dev_sync()
      !
      !DEV_ACC data present(X,Y)
      !DEV_ACC host_data use_device(X,Y)
#if defined __DXL_CUBLAS
      res= cublasZdotu(N,X,INCX,Y,INCY)
#elif defined __DXL_ROCBLAS
      !DEV_OMPGPU target data use_device_ptr(X,Y)
      ierr= rocblas_zdotu(handle,N,c_loc(X),INCX,c_loc(Y),INCY,c_loc(res))
      !DEV_OMPGPU end target data
#elif defined __DXL_MKL_GPU
      !DEV_OMPGPU target data use_device_addr(X,Y)
      !DEV_OMPGPU dispatch
      res= ZDOTU(N,X,INCX,Y,INCY)
      !DEV_OMPGPU end dispatch
      !DEV_OMPGPU end target data
#else
      res= ZDOTU(N,X,INCX,Y,INCY)
#endif
      !DEV_ACC end host_data
      !DEV_ACC end data
      !
      call acc_loc_dev_sync()
      !
   end function dev_ZDOTU_gpu
   !
   module function dev_CDOTC_gpu(N,X,INCX,Y,INCY) result(res)
      implicit none
      complex(real32) :: res
      integer,  intent(in) :: N
      integer,  intent(in) :: INCX,INCY
      complex(real32), intent(in) DEV_ATTR :: X(:)
      complex(real32)             DEV_ATTR :: Y(:)
      !
      integer :: ierr
      !
      call acc_loc_dev_sync()
      !
      !DEV_ACC data present(X,Y)
      !DEV_ACC host_data use_device(X,Y)
#if defined __DXL_CUBLAS
      res= cublasCdotc(N,X,INCX,Y,INCY)
#elif defined __DXL_ROCBLAS
      !DEV_OMPGPU target data use_device_ptr(X,Y)
      ierr= rocblas_cdotc(handle,N,c_loc(X),INCX,c_loc(Y),INCY,c_loc(res))
      !DEV_OMPGPU end target data
#elif defined __DXL_MKL_GPU
      !DEV_OMPGPU target data use_device_addr(X,Y)
      !DEV_OMPGPU dispatch
      res= CDOTC(N,X,INCX,Y,INCY)
      !DEV_OMPGPU end dispatch
      !DEV_OMPGPU end target data
#else
      res= CDOTC(N,X,INCX,Y,INCY)
#endif
      !DEV_ACC end host_data
      !DEV_ACC end data
      !
      call acc_loc_dev_sync()
      !
   end function dev_CDOTC_gpu
   !
   module function dev_ZDOTC_gpu(N,X,INCX,Y,INCY) result(res)
      implicit none
      complex(real64) :: res
      integer, intent(in) :: N
      integer, intent(in) :: INCX,INCY
      complex(real64), intent(in) DEV_ATTR :: X(:)
      complex(real64)             DEV_ATTR :: Y(:)
      !
      integer :: ierr
      !
      call acc_loc_dev_sync()
      !
      !DEV_ACC data present(X,Y)
      !DEV_ACC host_data use_device(X,Y)
#if defined __DXL_CUBLAS
      res= cublasZdotc(N,X,INCX,Y,INCY)
#elif defined __DXL_ROCBLAS
      !DEV_OMPGPU target data use_device_ptr(X,Y)
      ierr= rocblas_zdotc(handle,N,c_loc(X),INCX,c_loc(Y),INCY,c_loc(res))
      !DEV_OMPGPU end target data
#elif defined __DXL_MKL_GPU
      !DEV_OMPGPU target data use_device_addr(X,Y)
      !DEV_OMPGPU dispatch
      res= ZDOTC(N,X,INCX,Y,INCY)
      !DEV_OMPGPU end dispatch
      !DEV_OMPGPU end target data
#else
      res= ZDOTC(N,X,INCX,Y,INCY)
#endif
      !DEV_ACC end host_data
      !DEV_ACC end data
      !
      call acc_loc_dev_sync()
      !
   end function dev_ZDOTC_gpu
   !
end submodule devxlib_linalg_dot
