Home
last modified time | relevance | path

Searched refs:async (Results 1 – 25 of 287) sorted by relevance

12345678910>>...12

/llvm-project/mlir/test/Dialect/Async/
H A Dops.mlir4 func.func @identity_token(%arg0: !async.token) -> !async.token {
5 // CHECK: return %arg0 : !async.token
6 return %arg0 : !async.token
10 func.func @identity_value(%arg0 : !async.value<f32>) -> !async.value<f32> {
11 // CHECK: return %arg0 : !async.value<f32>
12 return %arg0 : !async.value<f32>
16 func.func @empty_async_execute() -> !async.token {
17 // CHECK: async.execute
18 %token = async.execute {
19 async.yield
[all …]
H A Druntime.mlir4 func.func @create_token() -> !async.token {
5 // CHECK: %0 = async.runtime.create : !async.token
6 %0 = async.runtime.create : !async.token
7 // CHECK: return %0 : !async.token
8 return %0 : !async.token
12 func.func @create_value() -> !async.value<f32> {
13 // CHECK: %0 = async.runtime.create : !async.value<f32>
14 %0 = async.runtime.create : !async.value<f32>
15 // CHECK: return %0 : !async.value<f32>
16 return %0 : !async.value<f32>
[all …]
H A Dasync-runtime-ref-counting.mlir1 // RUN: mlir-opt %s -async-runtime-ref-counting | FileCheck %s
4 func.func private @token() -> !async.token
10 func.func private @take_token(%arg0: !async.token)
13 // CHECK: %[[TOKEN:.*]]: !async.token
14 func.func @token_arg_no_uses(%arg0: !async.token) {
15 // CHECK: async.runtime.drop_ref %[[TOKEN]] {count = 1 : i64}
21 // CHECK: %[[TOKEN:.*]] = async.runtime.create : !async.token
22 // CHECK: async.runtime.drop_ref %[[TOKEN]] {count = 1 : i64}
23 %0 = async.runtime.create : !async.token
30 // CHECK: async.runtime.drop_ref %[[TOKEN]] {count = 1 : i64}
[all …]
H A Dasync-to-async-runtime.mlir1 // RUN: mlir-opt %s -split-input-file -async-func-to-async-runtime \
2 // RUN: -async-to-async-runtime | FileCheck %s --dump-input=always
6 %token = async.execute {
9 async.yield
11 async.await %token : !async.token
15 // Function outlined from the async.execute operation.
17 // CHECK-SAME: -> !async.token
20 // CHECK: %[[TOKEN:.*]] = async.runtime.create : !async.token
21 // CHECK: %[[ID:.*]] = async.coro.id
22 // CHECK: %[[HDL:.*]] = async.coro.begin
[all …]
H A Dasync-runtime-ref-counting-opt.mlir1 // RUN: mlir-opt %s -async-runtime-ref-counting-opt | FileCheck %s
3 func.func private @consume_token(%arg0: !async.token)
6 func.func @cancellable_operations_0(%arg0: !async.token) {
7 // CHECK-NOT: async.runtime.add_ref
8 // CHECK-NOT: async.runtime.drop_ref
9 async.runtime.add_ref %arg0 {count = 1 : i64} : !async.token
10 async.runtime.drop_ref %arg0 {count = 1 : i64} : !async.token
16 func.func @cancellable_operations_1(%arg0: !async.token) {
17 // CHECK-NOT: async.runtime.add_ref
18 async.runtime.add_ref %arg0 {count = 1 : i64} : !async.token
[all …]
H A Dasync-runtime-policy-based-ref-counting.mlir1 // RUN: mlir-opt %s -async-runtime-policy-based-ref-counting | FileCheck %s
4 // CHECK: %[[TOKEN:.*]]: !async.token
5 func.func @token_await(%arg0: !async.token) {
6 // CHECK: async.runtime.await %[[TOKEN]]
7 // CHECK-NOT: async.runtime.drop_ref
8 async.runtime.await %arg0 : !async.token
13 // CHECK: %[[GROUP:.*]]: !async.group
14 func.func @group_await(%arg0: !async.group) {
15 // CHECK: async.runtime.await %[[GROUP]]
16 // CHECK-NOT: async.runtime.drop_ref
[all …]
H A Dcoro.mlir4 func.func @coro_id() -> !async.coro.id {
5 // CHECK: %0 = async.coro.id
6 // CHECK: return %0 : !async.coro.id
7 %0 = async.coro.id
8 return %0 : !async.coro.id
12 func.func @coro_handle(%arg0: !async.coro.id) -> !async.coro.handle {
13 // CHECK: %0 = async.coro.begin %arg0
14 // CHECK: return %0 : !async.coro.handle
15 %0 = async.coro.begin %arg0
16 return %0 : !async.coro.handle
[all …]
H A Dverify.mlir5 func.func @no_op(%arg0: !async.token) {
12 …// expected-error @+1 {{'async.await' op operand #0 must be async value type or async token type, …
13 async.await %arg0 : f32
18 func.func @wrong_async_await_result_type(%arg0: !async.value<f32>) {
19 // expected-error @+1 {{'async.await' op result type 'f64' does not match async value type 'f32'}}
20 %0 = "async.await"(%arg0): (!async.value<f32>) -> f64
25 // expected-error @+1 {{'async.func' op result is expected to be at least of size 1, but got 0}}
26 async.func @wrong_async_func_void_result_type(%arg0: f32) {
32 // expected-error @+1 {{'async.func' op result type must be async value type or async token type, b…
33 async.func @wrong_async_func_result_type(%arg0: f32) -> f32 {
[all …]
/llvm-project/mlir/test/mlir-cpu-runner/
H A Dasync-func.mlir
H A Dasync-error.mlir
H A Dasync-group.mlir
H A Dasync-value.mlir
/llvm-project/mlir/test/Dialect/GPU/
H A Dasync-region.mlir1 // RUN: mlir-opt -gpu-async-region %s | FileCheck %s
12 // CHECK-LABEL:func @async(%{{.*}}: index)
13 func.func @async(%sz : index) {
14 // CHECK: %[[t0:.*]] = gpu.wait async
15 // CHECK: %[[t1:.*]] = gpu.launch_func async [%[[t0]]]
18 // CHECK: %[[t2:.*]] = gpu.launch_func async [%[[t1]]]
21 // CHECK: %[[m:.*]], %[[t3:.*]] = gpu.alloc async [%[[t2]]] ()
23 // CHECK: %[[t4:.*]] = gpu.dealloc async [%[[t3]]] %[[m]]
33 // CHECK: %[[a0:.*]], %[[f0:.*]] = async.execute
34 %a0 = async.execute {
[all …]
H A Dsparse-roundtrip.mlir6 // CHECK: %{{.*}} = gpu.wait async
7 // CHECK: %{{.*}}, %{{.*}} = gpu.alloc async [%{{.*}}] (%{{.*}}) : memref<?xindex>
8 // CHECK: %{{.*}}, %{{.*}} = gpu.alloc async [%{{.*}}] (%{{.*}}) : memref<?xf64>
9 …// CHECK: %{{.*}}, %{{.*}} = gpu.create_coo async [%{{.*}}] %{{.*}}, %{{.*}}, %{{.*}}, %{{.*}}, %{…
10 …// CHECK: %{{.*}}, %{{.*}} = gpu.create_dn_tensor async [%{{.*}}] %{{.*}}, %{{.*}} : index into me…
11 …// CHECK: %{{.*}}, %{{.*}} = gpu.spmv_buffer_size async [%{{.*}}] %{{.*}}, %{{.*}}, %{{.*}} into f…
12 …// CHECK: %{{.*}} = gpu.spmv async [%{{.*}}] %{{.*}}, %{{.*}}, %{{.*}}, %{{.*}} : memref<?xf64> in…
13 // CHECK: %{{.*}} = gpu.destroy_sp_mat async [%{{.*}}] %{{.*}}
14 // CHECK: %{{.*}} = gpu.destroy_dn_tensor async [%{{.*}}] %{{.*}}
18 %token0 = gpu.wait async
[all …]
/llvm-project/mlir/test/Conversion/AsyncToLLVM/
H A Dconvert-runtime-to-llvm.mlir1 // RUN: mlir-opt %s -convert-async-to-llvm | FileCheck %s --dump-input=always
6 %0 = async.runtime.create : !async.token
16 %0 = async.runtime.create : !async.value<f32>
26 %0 = async.runtime.create_group %c: !async.group
33 %0 = async.runtime.create : !async.token
35 async.runtime.set_available %0 : !async.token
42 %0 = async.runtime.create : !async.value<f32>
44 async.runtime.set_available %0 : !async.value<f32>
51 %0 = async.runtime.create : !async.token
53 %1 = async.runtime.is_error %0 : !async.token
[all …]
H A Dconvert-to-llvm.mlir1 // RUN: mlir-opt %s -split-input-file -async-to-async-runtime -convert-async-to-llvm | FileCheck %s
4 func.func @reference_counting(%arg0: !async.token) {
7 async.runtime.add_ref %arg0 {count = 2 : i64} : !async.token
11 async.runtime.drop_ref %arg0 {count = 1 : i64} : !async.token
21 %token = async.execute {
24 async.yield
32 async.await %token : !async.token
36 // Function outlined from the async.execute operation.
44 // Pass a suspended coroutine to the async runtime.
76 %token0 = async.execute {
[all …]
/llvm-project/mlir/test/Dialect/SparseTensor/GPU/
H A Dgpu_combi.mlir15 // CHECK: gpu.alloc async
16 // CHECK: gpu.memcpy async
17 // CHECK: gpu.alloc async
18 // CHECK: gpu.memcpy async
19 // CHECK: gpu.alloc async
20 // CHECK: gpu.memcpy async
21 // CHECK: gpu.alloc async
22 // CHECK: gpu.memcpy async
23 // CHECK: gpu.alloc async
24 // CHECK: gpu.memcpy async
[all …]
H A Dgpu_spgemm_lib.mlir19 // CHECK: %[[VAL_14:.*]] = gpu.wait async
21 // CHECK: %[[VAL_16:.*]], %[[VAL_17:.*]] = gpu.alloc async {{\[}}%[[VAL_14]]] (%[[VAL_15]…
22 // CHECK: %[[VAL_18:.*]] = gpu.memcpy async {{\[}}%[[VAL_17]]] %[[VAL_16]], %[[VAL_8]] : …
23 // CHECK: %[[VAL_19:.*]] = gpu.wait async
25 // CHECK: %[[VAL_21:.*]], %[[VAL_22:.*]] = gpu.alloc async {{\[}}%[[VAL_19]]] (%[[VAL_20]…
26 // CHECK: %[[VAL_23:.*]] = gpu.memcpy async {{\[}}%[[VAL_22]]] %[[VAL_21]], %[[VAL_9]] : …
27 // CHECK: %[[VAL_24:.*]] = gpu.wait async
29 // CHECK: %[[VAL_26:.*]], %[[VAL_27:.*]] = gpu.alloc async {{\[}}%[[VAL_24]]] (%[[VAL_25]…
30 // CHECK: %[[VAL_28:.*]] = gpu.memcpy async {{\[}}%[[VAL_27]]] %[[VAL_26]], %[[VAL_10]] :…
31 // CHECK: %[[VAL_29:.*]] = gpu.wait async
[all …]
H A Dgpu_matvec_lib.mlir21 // CHECK: %[[VAL_11:.*]] = gpu.wait async
23 // CHECK: %[[VAL_13:.*]], %[[VAL_14:.*]] = gpu.alloc async {{\[}}%[[VAL_11]]] (%[[VAL_12]]) : memref<?xindex>
24 // CHECK: %[[VAL_15:.*]] = gpu.memcpy async {{\[}}%[[VAL_14]]] %[[VAL_13]], %[[VAL_8]] : memref<?xindex>, memref<?xindex, strided<[?], offset: ?>>
25 // CHECK: %[[VAL_16:.*]] = gpu.wait async
27 // CHECK: %[[VAL_18:.*]], %[[VAL_19:.*]] = gpu.alloc async {{\[}}%[[VAL_16]]] (%[[VAL_17]]) : memref<?xindex>
28 // CHECK: %[[VAL_20:.*]] = gpu.memcpy async {{\[}}%[[VAL_19]]] %[[VAL_18]], %[[VAL_9]] : memref<?xindex>, memref<?xindex, strided<[?], offset: ?>>
29 // CHECK: %[[VAL_21:.*]] = gpu.wait async
31 // CHECK: %[[VAL_23:.*]], %[[VAL_24:.*]] = gpu.alloc async {{\[}}%[[VAL_21]]] (%[[VAL_22]]) : memref<?xf64>
32 // CHECK: %[[VAL_25:.*]] = gpu.memcpy async {{\[}}%[[VAL_24]]] %[[VAL_23]], %[[VAL_10]] : memref<?xf64>, memref<?xf64>
34 // CHECK: %[[VAL_27:.*]] = gpu.wait async
[all...]
/llvm-project/clang/test/SemaOpenACC/
H A Dcompute-construct-async-clause.cpp23 #pragma acc parallel async in Test()
25 #pragma acc parallel async(1) in Test()
27 #pragma acc kernels async(-51) in Test()
30 // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('struct NotConvertible' invalid}} in Test()
31 #pragma acc parallel async(NC) in Test()
36 #pragma acc kernels async(*SomeIncomplete) in Test()
39 #pragma acc parallel async(SomeE) in Test()
42 // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('enum E2' invalid}} in Test()
43 #pragma acc kernels async(SomeE2) in Test()
46 #pragma acc parallel async(Conver in Test()
[all...]
/llvm-project/mlir/include/mlir/Dialect/Async/
H A DPasses.td14 def AsyncParallelFor : Pass<"async-parallel-for", "ModuleOp"> {
15 let summary = "Convert scf.parallel operations to multiple async compute ops "
20 Option<"asyncDispatch", "async-dispatch",
22 "Dispatch async compute tasks using recursive work splitting. If `false` "
23 "async compute tasks will be launched using simple for loop in the "
28 "The number of available workers to execute async operations. If `-1` "
38 "async::AsyncDialect",
44 def AsyncToAsyncRuntime : Pass<"async-to-async-runtime", "ModuleOp"> {
45 let summary = "Lower all high level async operations (e.g. async.execute) to"
46 "the explicit async.runtime and async.coro operations";
[all …]
/llvm-project/llvm/test/CodeGen/NVPTX/
H A Dasync-copy.ll6 declare void @llvm.nvvm.cp.async.wait.group(i32)
10 ; CHECK: cp.async.wait_group 8;
11 tail call void @llvm.nvvm.cp.async.wait.group(i32 8)
12 ; CHECK: cp.async.wait_group 0;
13 tail call void @llvm.nvvm.cp.async.wait.group(i32 0)
14 ; CHECK: cp.async.wait_group 16;
15 tail call void @llvm.nvvm.cp.async.wait.group(i32 16)
19 declare void @llvm.nvvm.cp.async.wait.all()
23 ; CHECK: cp.async.wait_all
24 tail call void @llvm.nvvm.cp.async
[all...]
/llvm-project/mlir/include/mlir/Dialect/Async/IR/
H A DAsyncOps.td44 The `body` region attached to the `async.execute` operation semantically
54 state). All dependencies must be made explicit with async execute arguments
55 (`async.token` or `async.value`).
57 `async.execute` operation takes `async.token` dependencies and `async.value`
64 %dependency = ... : !async.token
65 %value = ... : !async.value<f32>
68 async
[all...]
/llvm-project/mlir/test/Conversion/GPUCommon/
H A Dlower-sparse-to-gpu-runtime-calls.mlir18 %token0 = gpu.wait async
19 %mem1, %token1 = gpu.alloc async [%token0] (%arg0) : memref<?xindex>
20 %mem2, %token2 = gpu.alloc async [%token1] (%arg0) : memref<?xf64>
21 …%spmat, %token4 = gpu.create_coo async [%token2] %arg0, %arg0, %arg0, %mem1, %mem1, %mem2 : memref…
22 %dnvec, %token5 = gpu.create_dn_tensor async [%token4] %mem2, %arg0 : index into memref<?xf64>
23 %bufferSz, %token6 = gpu.spmv_buffer_size async [%token5] %spmat, %dnvec, %dnvec into f64
24 %token7 = gpu.spmv async [%token6] %spmat, %dnvec, %dnvec, %mem2 : memref<?xf64> into f64
25 %token8 = gpu.destroy_sp_mat async [%token7] %spmat
26 %token9 = gpu.destroy_dn_tensor async [%token8] %dnvec
44 %token0 = gpu.wait async
[all …]
/llvm-project/llvm/test/Transforms/Coroutines/
H A Dcoro-async.ll5 %async.task = type { i64 }
6 %async.actor = type { i64 }
7 %async.fp = type <{ i32, i32 }>
9 %async.ctxt = type { ptr, ptr }
11 ; The async callee.
13 declare void @my_other_async_function(ptr %async.ctxt)
15 ; The current async function (the caller).
16 ; This struct describes an async function. The first field is the
17 ; relative offset to the async function implementation, the second field is the
18 ; size needed for the async contex
[all...]

12345678910>>...12