Added some modifications for GPU code

cmake/tests/test_GPU.cmake

@@ -8,6 +8,8 @@ if (CLI_FLEUR_USE_GPU)
       message("Using cuda8")
    elseif(${CLI_FLEUR_USE_GPU} MATCHES "cuda9")
       set(CMAKE_Fortran_FLAGS "${CMAKE_Fortran_FLAGS} -Mcuda=cuda9.0,cc60 -Mcuda=rdc -Mcudalib=cublas")
+   elseif(${CLI_FLEUR_USE_GPU} MATCHES "cuda9.1")
+      set(CMAKE_Fortran_FLAGS "${CMAKE_Fortran_FLAGS} -Mcuda=cuda9.1,cc60 -Mcuda=rdc -Mcudalib=cublas")
    elseif(${CLI_FLEUR_USE_GPU} MATCHES "nvtx")
       set(CMAKE_Fortran_FLAGS "${CMAKE_Fortran_FLAGS} -Mcuda=cuda9.0,cc60 -Mcuda=rdc -Mcudalib=cublas -lnvToolsExt ")
    elseif(${CLI_FLEUR_USE_GPU} MATCHES "emu")
@@ -21,7 +23,7 @@ if (CLI_FLEUR_USE_GPU)
 	    LINK_LIBRARIES "-lcusolver"
             )
    if (FLEUR_USE_CUSOLVER)
-     set(${FLEUR_LIBRARIES} "${FLEUR_LIBRARIES};-lcusolver")
+     set(FLEUR_LIBRARIES "${FLEUR_LIBRARIES};-lcusolver")
      set(FLEUR_MPI_DEFINITIONS ${FLEUR_MPI_DEFINITIONS} "CPP_CUSOLVER")
      set(FLEUR_DEFINITIONS ${FLEUR_DEFINITIONS} "CPP_CUSOLVER")
    endif()

eigen/eigen_redist_matrix.f90

@@ -38,22 +38,22 @@ CONTAINS
     !up-up component (or only component in collinear case)
     IF (SIZE(mat)==1) THEN
        CALL mat_final%move(mat(1,1))
-       CALL mat(1,1)%free()
+       !CALL mat(1,1)%free()
        RETURN
     ENDIF
 
     CALL mat_final%copy(mat(1,1),1,1)
-    CALL mat(1,1)%free()
+    !CALL mat(1,1)%free()
   
     !down-down component
     CALL mat_final%copy(mat(2,2),lapw%nv(1)+atoms%nlotot+1,lapw%nv(1)+atoms%nlotot+1)
-    CALL mat(2,2)%free()
+    !CALL mat(2,2)%free()
 
     !Now collect off-diagonal parts
     CALL mat(1,2)%add_transpose(mat(2,1))
     CALL mat_final%copy(mat(1,2),1,lapw%nv(1)+atoms%nlotot+1)
-    CALL mat(1,2)%free()
-    CALL mat(2,1)%free()
+    !CALL mat(1,2)%free()
+    !CALL mat(2,1)%free()
     
   END SUBROUTINE eigen_redist_matrix
 END MODULE m_eigen_redist_matrix

types/types_gpumat.F90

 MODULE m_types_gpumat
   USE m_judft
   USE m_types_mat
+  USE cudafor
   IMPLICIT NONE
 
   !<Some routines are overwritten for GPU handling
@@ -102,6 +103,7 @@ CONTAINS
              CALL judft_error("Inconsistency between data types in m_mat")
           END IF
        END SELECT
+       i=cudaDeviceSynchronize()
     END SELECT
   END SUBROUTINE t_gpumat_add_transpose
 
@@ -159,6 +161,8 @@ CONTAINS
              ENDDO
           ENDIF
        END SELECT
+
+       i=cudaDeviceSynchronize()
     END IF
   end SUBROUTINE t_gpumat_transpose
 
@@ -250,6 +254,7 @@ CONTAINS
                 ENDDO
              ENDDO
           END IF
+          i=cudaDeviceSynchronize()
           CLASS default
           no_gpu=.TRUE.
        END SELECT
@@ -294,6 +299,7 @@ CONTAINS
              ENDDO
           ENDDO
        ENDIF
+       n=cudaDeviceSynchronize()
     END SELECT
   END SUBROUTINE t_gpumat_clear
 #endif  

types/types_mat.F90

@@ -82,23 +82,18 @@ MODULE m_types_mat
      IMPLICIT NONE
      CLASS(t_mat),INTENT(INOUT) :: mat
      CLASS(t_mat),INTENT(IN)    :: templ
-     SELECT TYPE(templ)
-     TYPE is(t_mat)
-        mat%l_real=templ%l_real
-        mat%matsize1=templ%matsize1
-        mat%matsize2=templ%matsize2
-        IF (mat%l_real) THEN
-           ALLOCATE(mat%data_r(mat%matsize1,mat%matsize2))
-           ALLOCATE(mat%data_c(1,1))
-           mat%data_r=0.0
-        ELSE
-           ALLOCATE(mat%data_c(mat%matsize1,mat%matsize2))
-           ALLOCATE(mat%data_r(1,1))
-           mat%data_c=0.0
-        END IF
-     CLASS default
-        CALL judft_error("Mixed initialization in t_mat not possible(BUG)")
-     END SELECT
+     mat%l_real=templ%l_real
+     mat%matsize1=templ%matsize1
+     mat%matsize2=templ%matsize2
+     IF (mat%l_real) THEN
+        ALLOCATE(mat%data_r(mat%matsize1,mat%matsize2))
+        ALLOCATE(mat%data_c(1,1))
+        mat%data_r=0.0
+     ELSE
+        ALLOCATE(mat%data_c(mat%matsize1,mat%matsize2))
+        ALLOCATE(mat%data_r(1,1))
+        mat%data_c=0.0
+     END IF
    END SUBROUTINE t_mat_init_template
      
   SUBROUTINE t_mat_alloc(mat,l_real,matsize1,matsize2,init)