Drop support for CUDA 10.1 and below.

Author: maleadt

.buildkite/pipeline.yml

@@ -260,27 +260,6 @@ steps:
     if: build.message !~ /\[skip tests\]/ && !build.pull_request.draft
     timeout_in_minutes: 120
 
-  - label: "CUDA 10.1"
-    plugins:
-      - JuliaCI/julia#v1:
-          version: 1.6
-      - JuliaCI/julia-test#v1: ~
-      - JuliaCI/julia-coverage#v1:
-          codecov: true
-          dirs:
-            - src
-            - lib
-            - examples
-    agents:
-      queue: "juliagpu"
-      cuda: "10.1"
-      cap: "sm_75"
-    env:
-      JULIA_CUDA_VERSION: '10.1'
-      JULIA_CUDA_USE_BINARYBUILDER: 'true'
-    if: build.message !~ /\[skip tests\]/ && !build.pull_request.draft
-    timeout_in_minutes: 120
-
 
    # special tests
 

lib/cublas/wrappers.jl

@@ -103,17 +103,9 @@ for (fname, elty) in ((:cublasDscal_v2,:Float64),
     end
 end
 function scal!(n::Integer, alpha::Number, x::StridedCuArray{Float16})
-    if version() > v"10.1"
-        α = convert(Float32, alpha)
-        cublasScalEx(handle(), n, Ref{Float32}(α), Float32, x, Float16, stride(x, 1), Float32)
-        return x
-    else
-        wide_x = widen.(x)
-        scal!(n, alpha, wide_x)
-        thin_x = convert(typeof(x), wide_x)
-        copyto!(x, thin_x)
-        return x
-    end
+    α = convert(Float32, alpha)
+    cublasScalEx(handle(), n, Ref{Float32}(α), Float32, x, Float16, stride(x, 1), Float32)
+    return x
 end
 # specific variants in case x is complex and alpha is real
 for (fname, elty, celty) in ((:cublasCsscal_v2, :Float32, :ComplexF32),
@@ -153,13 +145,9 @@ for (jname, fname, elty) in ((:dot,:cublasDdot_v2,:Float64),
     end
 end
 function dot(n::Integer, x::StridedCuArray{Float16}, y::StridedCuArray{Float16})
-    if version() > v"10.1"
-        result = Ref{Float16}()
-        cublasDotEx(handle(), n, x, Float16, stride(x, 1), y, Float16, stride(y, 1), result, Float16, Float32)
-        return result[]
-    else
-        return convert(Float16, dot(n, convert(CuArray{Float32}, x), convert(CuArray{Float32}, y)))
-    end
+    result = Ref{Float16}()
+    cublasDotEx(handle(), n, x, Float16, stride(x, 1), y, Float16, stride(y, 1), result, Float16, Float32)
+    return result[]
 end
 function dotc(n::Integer, x::StridedCuArray{ComplexF16}, y::StridedCuArray{ComplexF16})
     return convert(ComplexF16, dotc(n, convert(CuArray{ComplexF32}, x), convert(CuArray{ComplexF32}, y)))
@@ -185,15 +173,9 @@ end
 nrm2(x::StridedCuArray) = nrm2(length(x), x)
 
 function nrm2(n::Integer, x::StridedCuArray{Float16})
-    if version() > v"10.1"
-        result = Ref{Float16}()
-        cublasNrm2Ex(handle(), n, x, Float16, stride(x, 1), result, Float16, Float32)
-        return result[]
-    else
-        wide_x = widen.(x)
-        nrm    = nrm2(n, wide_x)
-        return convert(Float16, nrm)
-    end
+    result = Ref{Float16}()
+    cublasNrm2Ex(handle(), n, x, Float16, stride(x, 1), result, Float16, Float32)
+    return result[]
 end
 function nrm2(n::Integer, x::StridedCuArray{ComplexF16})
     wide_x = widen.(x)
@@ -233,18 +215,9 @@ for (fname, elty) in ((:cublasDaxpy_v2,:Float64),
 end
 
 function axpy!(n::Integer, alpha::Number, dx::StridedCuArray{Float16}, dy::StridedCuArray{Float16})
-    if version() >= v"10.1"
-        α = convert(Float32, alpha)
-        cublasAxpyEx(handle(), n, Ref{Float32}(α), Float32, dx, Float16, stride(dx, 1), dy, Float16, stride(dy, 1), Float32)
-        return dy
-    else
-        wide_x = widen.(dx)
-        wide_y = widen.(dy)
-        axpy!(n, alpha, wide_x, wide_y)
-        thin_y = convert(typeof(dy), wide_y)
-        copyto!(dy, thin_y)
-        return dy
-    end
+    α = convert(Float32, alpha)
+    cublasAxpyEx(handle(), n, Ref{Float32}(α), Float32, dx, Float16, stride(dx, 1), dy, Float16, stride(dy, 1), Float32)
+    return dy
 end
 function axpy!(n::Integer, alpha::Number, dx::StridedCuArray{ComplexF16}, dy::StridedCuArray{ComplexF16})
     wide_x = widen.(dx)

lib/cusparse/conversions.jl

@@ -121,36 +121,27 @@ function CuSparseMatrixCSR{T}(S::Adjoint{T, <:CuSparseMatrixCSC{T}}) where {T <:
 end
 
 # by flipping rows and columns, we can use that to get CSC to CSR too
-for (fname,elty) in ((:cusparseScsr2csc, :Float32),
-                     (:cusparseDcsr2csc, :Float64),
-                     (:cusparseCcsr2csc, :ComplexF32),
-                     (:cusparseZcsr2csc, :ComplexF64))
+for elty in (Float32, Float64, ComplexF32, ComplexF64)
     @eval begin
         function CuSparseMatrixCSC{$elty}(csr::CuSparseMatrixCSR{$elty}; inda::SparseChar='O')
             m,n = size(csr)
             colPtr = CUDA.zeros(Cint, n+1)
             rowVal = CUDA.zeros(Cint, nnz(csr))
             nzVal = CUDA.zeros($elty, nnz(csr))
-            if version() >= v"10.2"
-                # TODO: algorithm configuratibility?
-                function bufferSize()
-                    out = Ref{Csize_t}(1)
-                    cusparseCsr2cscEx2_bufferSize(handle(), m, n, nnz(csr), nonzeros(csr),
-                        csr.rowPtr, csr.colVal, nzVal, colPtr, rowVal,
-                        $elty, CUSPARSE_ACTION_NUMERIC, inda,
-                        CUSPARSE_CSR2CSC_ALG1, out)
-                    return out[]
-                end
-                with_workspace(bufferSize) do buffer
-                    cusparseCsr2cscEx2(handle(), m, n, nnz(csr), nonzeros(csr),
-                        csr.rowPtr, csr.colVal, nzVal, colPtr, rowVal,
-                        $elty, CUSPARSE_ACTION_NUMERIC, inda,
-                        CUSPARSE_CSR2CSC_ALG1, buffer)
-                end
-            else
-                $fname(handle(), m, n, nnz(csr), nonzeros(csr),
-                    csr.rowPtr, csr.colVal, nzVal, rowVal,
-                    colPtr, CUSPARSE_ACTION_NUMERIC, inda)
+            # TODO: algorithm configuratibility?
+            function bufferSize()
+                out = Ref{Csize_t}(1)
+                cusparseCsr2cscEx2_bufferSize(handle(), m, n, nnz(csr), nonzeros(csr),
+                    csr.rowPtr, csr.colVal, nzVal, colPtr, rowVal,
+                    $elty, CUSPARSE_ACTION_NUMERIC, inda,
+                    CUSPARSE_CSR2CSC_ALG1, out)
+                return out[]
+            end
+            with_workspace(bufferSize) do buffer
+                cusparseCsr2cscEx2(handle(), m, n, nnz(csr), nonzeros(csr),
+                    csr.rowPtr, csr.colVal, nzVal, colPtr, rowVal,
+                    $elty, CUSPARSE_ACTION_NUMERIC, inda,
+                    CUSPARSE_CSR2CSC_ALG1, buffer)
             end
             CuSparseMatrixCSC(colPtr,rowVal,nzVal,size(csr))
         end
@@ -160,26 +151,20 @@ for (fname,elty) in ((:cusparseScsr2csc, :Float32),
             rowPtr = CUDA.zeros(Cint,m+1)
             colVal = CUDA.zeros(Cint,nnz(csc))
             nzVal  = CUDA.zeros($elty,nnz(csc))
-            if version() >= v"10.2"
-                # TODO: algorithm configuratibility?
-                function bufferSize()
-                    out = Ref{Csize_t}(1)
-                    cusparseCsr2cscEx2_bufferSize(handle(), n, m, nnz(csc), nonzeros(csc),
-                        csc.colPtr, rowvals(csc), nzVal, rowPtr, colVal,
-                        $elty, CUSPARSE_ACTION_NUMERIC, inda,
-                        CUSPARSE_CSR2CSC_ALG1, out)
-                    return out[]
-                end
-                with_workspace(bufferSize) do buffer
-                    cusparseCsr2cscEx2(handle(), n, m, nnz(csc), nonzeros(csc),
-                        csc.colPtr, rowvals(csc), nzVal, rowPtr, colVal,
-                        $elty, CUSPARSE_ACTION_NUMERIC, inda,
-                        CUSPARSE_CSR2CSC_ALG1, buffer)
-                end
-            else
-                $fname(handle(), n, m, nnz(csc), nonzeros(csc),
-                    csc.colPtr, rowvals(csc), nzVal, colVal,
-                    rowPtr, CUSPARSE_ACTION_NUMERIC, inda)
+            # TODO: algorithm configuratibility?
+            function bufferSize()
+                out = Ref{Csize_t}(1)
+                cusparseCsr2cscEx2_bufferSize(handle(), n, m, nnz(csc), nonzeros(csc),
+                    csc.colPtr, rowvals(csc), nzVal, rowPtr, colVal,
+                    $elty, CUSPARSE_ACTION_NUMERIC, inda,
+                    CUSPARSE_CSR2CSC_ALG1, out)
+                return out[]
+            end
+            with_workspace(bufferSize) do buffer
+                cusparseCsr2cscEx2(handle(), n, m, nnz(csc), nonzeros(csc),
+                    csc.colPtr, rowvals(csc), nzVal, rowPtr, colVal,
+                    $elty, CUSPARSE_ACTION_NUMERIC, inda,
+                    CUSPARSE_CSR2CSC_ALG1, buffer)
             end
             CuSparseMatrixCSR(rowPtr,colVal,nzVal,size(csc))
         end
@@ -197,7 +182,7 @@ for (elty, welty) in ((:Float16, :Float32),
             rowVal = CUDA.zeros(Cint, nnz(csr))
             nzVal = CUDA.zeros($elty, nnz(csr))
             # TODO: algorithm configuratibility?
-            if version() >= v"10.2" && $elty == Float16 #broken for ComplexF16?
+            if $elty == Float16 #broken for ComplexF16?
                 function bufferSize()
                     out = Ref{Csize_t}(1)
                     cusparseCsr2cscEx2_bufferSize(handle(), m, n, nnz(csr), nonzeros(csr),
@@ -225,7 +210,7 @@ for (elty, welty) in ((:Float16, :Float32),
             rowPtr = CUDA.zeros(Cint,m+1)
             colVal = CUDA.zeros(Cint,nnz(csc))
             nzVal  = CUDA.zeros($elty,nnz(csc))
-            if version() >= v"10.2" && $elty == Float16 #broken for ComplexF16?
+            if $elty == Float16 #broken for ComplexF16?
                 # TODO: algorithm configuratibility?
                 function bufferSize()
                     out = Ref{Csize_t}(1)

src/initialization.jl

@@ -62,8 +62,8 @@ end
         return
     end
 
-    if version() < v"10.1"
-        @warn "This version of CUDA.jl only supports NVIDIA drivers for CUDA 10.1 or higher (yours is for CUDA $(version()))"
+    if version() < v"10.2"
+        @warn "This version of CUDA.jl only supports NVIDIA drivers for CUDA 10.2 or higher (yours is for CUDA $(version()))"
     end
 
     if version() < v"11.2"