Rename KernelIntrinsics module to KernelInterface

examples/histogram.jl

@@ -1,7 +1,7 @@
 # INCLUDE ROCM
 using KernelAbstractions, Test
 using KernelAbstractions: @atomic, @atomicswap, @atomicreplace
-import KernelAbstractions.KernelIntrinsics as KI
+import KernelAbstractions.KernelInterface as KI
 
 include(joinpath(dirname(pathof(KernelAbstractions)), "../examples/utils.jl")) # Load backend
 

examples/performant_matmul.jl

@@ -1,5 +1,5 @@
 using KernelAbstractions
-import KernelAbstractions.KernelIntrinsics as KI
+import KernelAbstractions.KernelInterface as KI
 
 using StaticArrays
 using Test

src/KernelAbstractions.jl

@@ -199,9 +199,9 @@ Abstract type for all KernelAbstractions backends.
 """
 abstract type Backend end
 
-include("intrinsics.jl")
-import .KernelIntrinsics as KI
-export KernelIntrinsics
+include("interface.jl")
+import .KernelInterface as KI
+export KernelInterface
 
 ###
 # Kernel language

src/intrinsics.jl → src/interface.jl

@@ -1,16 +1,16 @@
 """
-# `KernelIntrinsics`
+# `KernelInterface`
 
-The `KernelIntrinsics` (or `KI`) module defines the API interface for backends to define various lower-level device and
-host-side functionality. The `KI` intrinsics are used to define the higher-level device-side
-intrinsics functionality in `KernelAbstractions`.
+The `KernelInterface` (or `KI`) module defines the API interface for backends to define various lower-level device and
+host-side functionality. The `KI` interface is used to define the higher-level device-side
+functionality in `KernelAbstractions`.
 
 Both provide APIs for host and device-side functionality, but `KI` focuses on on lower-level
 functionality that is shared amongst backends, while `KernelAbstractions` provides higher-level functionality
 such as writing kernels that work on arrays with an arbitrary number of dimensions, or convenience functions
 like allocating arrays on a backend.
 """
-module KernelIntrinsics
+module KernelInterface
 
 import ..KernelAbstractions: Backend
 import GPUCompiler: split_kwargs, assign_args!

src/pocl/backend.jl

@@ -5,7 +5,7 @@ using ..POCL: @device_override, cl, method_table
 using ..POCL: device, clconvert, clfunction
 
 import KernelAbstractions as KA
-import KernelAbstractions.KernelIntrinsics as KI
+import KernelAbstractions.KernelInterface as KI
 
 import StaticArrays
 

test/intrinsics.jl → test/interface.jl

@@ -1,4 +1,4 @@
-import KernelAbstractions.KernelIntrinsics as KI
+import KernelAbstractions.KernelInterface as KI
 
 struct KernelData
     global_size::Int
@@ -8,7 +8,7 @@ struct KernelData
     num_groups::Int
     group_id::Int
 end
-function test_intrinsics_kernel(results)
+function test_interface_kernel(results)
     i = KI.get_global_id().x
 
     if i <= length(results)
@@ -24,8 +24,8 @@ function test_intrinsics_kernel(results)
     return
 end
 
-function intrinsics_testsuite(backend, AT)
-    @testset "KernelIntrinsics Tests" begin
+function interface_testsuite(backend, AT)
+    @testset "KernelInterface Tests" begin
         @testset "Launch parameters" begin
             # 1d
             function launch_kernel1d(arr)
@@ -80,7 +80,7 @@ function intrinsics_testsuite(backend, AT)
             @test_throws ArgumentError (KI.@kernel backend() numworkgroups = (2, 2, 2) workgroupsize = (2, 2, 2, 2) launch_kernel3d(arr3d))
         end
 
-        @testset "Basic intrinsics functionality" begin
+        @testset "Basic interface functionality" begin
 
             @test KI.max_work_group_size(backend()) isa Int
             @test KI.multiprocessor_count(backend()) isa Int
@@ -90,7 +90,7 @@ function intrinsics_testsuite(backend, AT)
             numworkgroups = 4
             N = workgroupsize * numworkgroups
             results = AT(Vector{KernelData}(undef, N))
-            kernel = KI.@kernel backend() launch = false test_intrinsics_kernel(results)
+            kernel = KI.@kernel backend() launch = false test_interface_kernel(results)
 
             @test KI.kernel_max_work_group_size(kernel) isa Int
             @test KI.kernel_max_work_group_size(kernel; max_work_items = 1) == 1

test/testsuite.jl

@@ -1,7 +1,7 @@
 module Testsuite
 
 using ..KernelAbstractions
-import ..KernelAbstractions.KernelIntrinsics as KI
+import ..KernelAbstractions.KernelInterface as KI
 using ..Test
 
 # We can't add test-dependencies withouth breaking backend packages
@@ -27,7 +27,7 @@ end
 
 
 include("test.jl")
-include("intrinsics.jl")
+include("interface.jl")
 include("localmem.jl")
 include("private.jl")
 include("unroll.jl")
@@ -50,8 +50,8 @@ function testsuite(backend, backend_str, backend_mod, AT, DAT; skip_tests = Set{
         specialfunctions_testsuite(backend)
     end
 
-    @conditional_testset "Intrinsics" skip_tests begin
-        intrinsics_testsuite(backend, AT)
+    @conditional_testset "Interface" skip_tests begin
+        interface_testsuite(backend, AT)
     end
 
     @conditional_testset "Localmem" skip_tests begin