Merge pull request #826

Coverage for Strided Arrays for syrk, trsm, trmm and potrf

Author: luraess

src/blas/wrappers.jl

@@ -814,7 +814,7 @@ for (fname, elty) in ((:rocblas_dsyrk,:Float64),
     @eval begin
         function syrk!(
             uplo::Char, trans::Char, alpha::($elty),
-            A::ROCVecOrMat{$elty}, beta::($elty), C::ROCMatrix{$elty},
+            A::StridedROCVecOrMat{$elty}, beta::($elty), C::StridedROCMatrix{$elty},
         )
             mC, n = size(C)
             if mC != n throw(DimensionMismatch("C must be square")) end
@@ -979,7 +979,7 @@ for (mmname, smname, elty) in
     @eval begin
         function trmm!(
             side::Char, uplo::Char, transa::Char, diag::Char, alpha::($elty),
-            A::ROCMatrix{$elty}, B::ROCMatrix{$elty}, C::ROCMatrix{$elty},
+            A::StridedROCMatrix{$elty}, B::StridedROCMatrix{$elty}, C::StridedROCMatrix{$elty},
         )
             m, n = size(B)
             mA, nA = size(A)
@@ -997,13 +997,13 @@ for (mmname, smname, elty) in
         end
         function trmm(
             side::Char, uplo::Char, transa::Char, diag::Char, alpha::($elty),
-            A::ROCMatrix{$elty}, B::ROCMatrix{$elty},
+            A::StridedROCMatrix{$elty}, B::StridedROCMatrix{$elty},
         )
             trmm!(side, uplo, transa, diag, alpha, A, B, similar(B))
         end
         function trsm!(
             side::Char, uplo::Char, transa::Char, diag::Char, alpha::($elty),
-            A::ROCMatrix{$elty}, B::ROCMatrix{$elty},
+            A::StridedROCMatrix{$elty}, B::StridedROCMatrix{$elty},
         )
             m, n = size(B)
             mA, nA = size(A)
@@ -1018,7 +1018,7 @@ for (mmname, smname, elty) in
         end
         function trsm(
             side::Char, uplo::Char, transa::Char, diag::Char, alpha::($elty),
-            A::ROCMatrix{$elty}, B::ROCMatrix{$elty},
+            A::StridedROCMatrix{$elty}, B::StridedROCMatrix{$elty},
         )
             trsm!(side, uplo, transa, diag, alpha, A, copy(B))
         end

src/solver/highlevel.jl

@@ -5,7 +5,7 @@ for (fname, elty) in (
     (:rocsolver_zpotrf, :ComplexF64),
 )
     @eval begin
-        function potrf!(uplo::Char, A::ROCMatrix{$elty})
+        function potrf!(uplo::Char, A::StridedROCMatrix{$elty})
             chkuplo(uplo)
             n = checksquare(A)
             lda = max(1, stride(A, 2))
@@ -630,7 +630,7 @@ end
 
 for elty in (:Float32, :Float64, :ComplexF32, :ComplexF64)
     @eval begin
-        LinearAlgebra.LAPACK.potrf!(uplo::Char, A::ROCMatrix{$elty}) = rocSOLVER.potrf!(uplo, A)
+        LinearAlgebra.LAPACK.potrf!(uplo::Char, A::StridedROCMatrix{$elty}) = rocSOLVER.potrf!(uplo, A)
         LinearAlgebra.LAPACK.potrs!(uplo::Char, A::ROCMatrix{$elty}, B::ROCVecOrMat{$elty}) = rocSOLVER.potrs!(uplo, A, B)
         LinearAlgebra.LAPACK.sytrf!(uplo::Char, A::ROCMatrix{$elty}) = rocSOLVER.sytrf!(uplo, A)
         LinearAlgebra.LAPACK.sytrf!(uplo::Char, A::ROCMatrix{$elty}, ipiv::ROCVector{Cint}) = rocSOLVER.sytrf!(uplo, A, ipiv)

test/rocarray/blas.jl

@@ -404,6 +404,15 @@ end
             dC = rocBLAS.trsm('L', 'U', 'N', 'N', one(T), dA, dB)
             @test collect(dC) ≈ triu(A) \ B
         end
+        @testset "trsm strided" begin
+            A  = rand(T, 2m, 2m)
+            dA = view(ROCArray(A),1:m,1:m)
+            B  = rand(T, 2m, 2m)
+            dB = view(ROCArray(B),1:m,1:m)
+            dC = rocBLAS.trsm('L', 'U', 'N', 'N', one(T), dA, dB)
+            @test collect(dC) ≈ triu(view(A,1:m,1:m)) \ view(B,1:m,1:m)
+        end
+            
         @testset "trsm_batched" begin
             batch_count = 3
             A  = [rand(T, m, m) for ix in 1:batch_count]
@@ -441,6 +450,15 @@ end
             @test collect(dC) ≈ triu(A) * B
         end
 
+         @testset "trmm strided" begin
+            A  = rand(T, 2m, 2m)
+            dA = view(ROCArray(A),1:m,1:m)
+            B  = rand(T, 2m, 2m)
+            dB = view(ROCArray(B),1:m,1:m)
+            dC = rocBLAS.trmm('L', 'U', 'N', 'N', one(T), dA, dB)
+            @test collect(dC) ≈ triu(view(A,1:m,1:m)) * view(B,1:m,1:m)
+        end
+
         @testset "triangular-triangular mul"  for (TRa, ta, TRb, tb) in (
             (UpperTriangular, identity,  LowerTriangular, identity),
             (LowerTriangular, identity,  UpperTriangular, identity),
@@ -578,6 +596,26 @@ end
         h_C = triu(h_C)
         @test C ≈ h_C
     end
+    @testset "syrk strided T=$T" for T in (Float32, Float64, ComplexF32, ComplexF64)
+        # generate parameters
+        α = rand(T)
+        β = rand(T)
+        A = rand(T, 2m, 2m)
+        Abad = rand(T, 2m + 1, 2m + 1)
+        C = rand(T, 2m, 2m)
+        # move to device
+        d_A, d_Abad = ROCArray(A), ROCArray(Abad)
+        C   = C + transpose(C)
+        d_C = ROCArray(C)
+        A_view = view(A,1:m,1:m)
+        C   = α*(A_view*transpose(A_view)) + β*view(C,1:m,1:m)
+        rocBLAS.syrk!('U','N',α,view(d_A,1:m,1:m),β,view(d_C,1:m,1:m))
+        # move back to host and compare
+        C   = triu(C)
+        h_C = Array(view(d_C,1:m,1:m))
+        h_C = triu(h_C)
+        @test C ≈ h_C
+    end
     @testset "syr2k T=$T" for T in (Float32, Float64, ComplexF32, ComplexF64)
         # generate parameters
         α = rand(T)

test/rocarray/solver.jl

@@ -183,6 +183,21 @@ end
         LAPACK.potrs!('U',A,B)
         @test B ≈ collect(d_B)
     end
+    @testset "elty = $elty strided" for elty in [Float32, Float64, ComplexF32, ComplexF64]
+        A    = rand(elty,n*2,n*2)
+        A    = A*A' + I
+        B    = rand(elty,n,p)
+        d_A  = view(ROCArray(A),1:n,1:n)
+        d_B  = ROCArray(B)
+
+        LAPACK.potrf!('L',d_A)
+        LAPACK.potrs!('U',copy(d_A),d_B)
+        LAPACK.potrf!('L',view(A,1:n,1:n))
+        LAPACK.potrs!('U',copy(view(A,1:n,1:n)),B)
+        @test B ≈ collect(d_B)
+    end
+
+        
 end
 
 @testset "sytrf!" begin