From 830762b85150595555b60db8876db54fa795e0cf Mon Sep 17 00:00:00 2001 From: Robert Imschweiler Date: Mon, 27 Oct 2025 10:09:22 -0500 Subject: [PATCH 1/2] [smoke-limbo] Remove unnecessary map order requirements --- .../usm_locals.cpp | 16 ++++++++-------- .../usm_locals.cpp | 16 ++++++++-------- .../usm_locals.cpp | 16 ++++++++-------- .../usm_locals.cpp | 14 +++++++------- .../usm_locals.cpp | 14 +++++++------- 5 files changed, 38 insertions(+), 38 deletions(-) diff --git a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp index f29aa6527..11ba345bd 100644 --- a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp @@ -16,10 +16,10 @@ int main() { for(size_t t = 0; t < 5; t++) k[t] = -t; - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) #pragma omp target update to(z[:10]) #pragma omp target map(tofrom:k[:5]) map(always, tofrom:x) map(tofrom:y) @@ -30,10 +30,10 @@ int main() { z[t]++; printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) #pragma omp target update from(z[:10]) // Note: when the output is redirected rather than printed at the console, diff --git a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp index 1901a9af5..f6d7a6399 100644 --- a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp @@ -16,10 +16,10 @@ int main() { for(size_t t = 0; t < 5; t++) k[t] = -t; - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) #pragma omp target update to(z[:10]) #pragma omp target map(tofrom:k[:5]) map(always, tofrom:x) map(tofrom:y) @@ -30,10 +30,10 @@ int main() { z[t]++; printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) #pragma omp target update from(z[:10]) // Note: when the output is redirected rather than printed at the console, diff --git a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp index 72b1f788c..bf7d1b758 100644 --- a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp @@ -16,10 +16,10 @@ int main() { for(size_t t = 0; t < 5; t++) k[t] = -t; - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) #pragma omp target update to(z[:10]) #pragma omp target map(tofrom:k[:5]) map(always, tofrom:x) map(tofrom:y) @@ -30,10 +30,10 @@ int main() { z[t]++; printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) #pragma omp target update from(z[:10]) // Note: when the output is redirected rather than printed at the console, diff --git a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp index 3a1899a5b..275e3af26 100644 --- a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp @@ -16,10 +16,10 @@ int main() { for(size_t t = 0; t < 5; t++) k[t] = -t; - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) #pragma omp target update to(z[:10]) #pragma omp target map(to:k[:5]) map(always, tofrom:x) map(tofrom:y) @@ -32,9 +32,9 @@ int main() { printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } #pragma omp target update from(z[:10]) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) // Note: when the output is redirected rather than printed at the console, // the printf'd strings are printed AFTER all the OpenMP runtime library diff --git a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp index 18b61889b..112ccb486 100644 --- a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp @@ -16,10 +16,10 @@ int main() { for(size_t t = 0; t < 5; t++) k[t] = -t; - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_submit_async: {{.*}} 0 ({{.*}} 20, {{.*}}) #pragma omp target update to(z[:10]) #pragma omp target map(to:k[:5]) map(always, tofrom:x) map(tofrom:y) @@ -31,9 +31,9 @@ int main() { printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } #pragma omp target update from(z[:10]) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) - /// CHECK: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) + /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) // Note: when the output is redirected rather than printed at the console, // the printf'd strings are printed AFTER all the OpenMP runtime library From 8ca2c3446d0f6e85113954b3cdfeb836a01d9318 Mon Sep 17 00:00:00 2001 From: Robert Imschweiler Date: Mon, 27 Oct 2025 10:23:36 -0500 Subject: [PATCH 2/2] separate submit DAG from retrieve DAG --- .../usm_locals.cpp | 3 +++ .../usm-locals-xnack-disabled-xnack-any/usm_locals.cpp | 3 +++ .../usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp | 3 +++ .../usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp | 3 +++ .../usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp | 3 +++ 5 files changed, 15 insertions(+) diff --git a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp index 11ba345bd..afb629436 100644 --- a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any-apu-maps/usm_locals.cpp @@ -30,6 +30,9 @@ int main() { z[t]++; printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } + // Note separate the two DAGs to ensure that retrieving is happening after + // submitting. + /// CHECK: launch_kernel: /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) diff --git a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp index f6d7a6399..daf36c7ef 100644 --- a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-any/usm_locals.cpp @@ -30,6 +30,9 @@ int main() { z[t]++; printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } + // Note separate the two DAGs to ensure that retrieving is happening after + // submitting. + /// CHECK: launch_kernel: /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) diff --git a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp index bf7d1b758..858e5b0db 100644 --- a/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-disabled-xnack-minus/usm_locals.cpp @@ -30,6 +30,9 @@ int main() { z[t]++; printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } + // Note separate the two DAGs to ensure that retrieving is happening after + // submitting. + /// CHECK: launch_kernel: /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 20, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) diff --git a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp index 275e3af26..bff625d92 100644 --- a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-any-non-apu/usm_locals.cpp @@ -32,6 +32,9 @@ int main() { printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } #pragma omp target update from(z[:10]) + // Note separate the two DAGs to ensure that retrieving is happening after + // submitting. + /// CHECK: launch_kernel: /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) diff --git a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp index 112ccb486..1ae59f42c 100644 --- a/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp +++ b/test/smoke-limbo/usm-locals-xnack-enabled-xnack-plus-non-apu/usm_locals.cpp @@ -31,6 +31,9 @@ int main() { printf("Device pointer for k = %p, k[3] = %d\n", k, k[3]); } #pragma omp target update from(z[:10]) + // Note separate the two DAGs to ensure that retrieving is happening after + // submitting. + /// CHECK: launch_kernel: /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 40, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}}) /// CHECK-DAG: data_retrieve_async: {{.*}} 0 ({{.*}} 4, {{.*}})