[llvm] [Offload][WIP] Move `/openmp/libomptarget` to `/offload` (PR #75125)
via llvm-commits
llvm-commits at lists.llvm.org
Mon Dec 11 17:35:09 PST 2023
github-actions[bot] wrote:
<!--LLVM CODE FORMAT COMMENT: {clang-format}-->
:warning: C/C++ code formatter, clang-format found issues in your code. :warning:
<details>
<summary>
You can test this locally with the following command:
</summary>
``````````bash
git-clang-format --diff 5dd1fc700857eb46882c7e27cd8835ef4fee1f22 e8b31e143337325ff46422bb62e5ead445868a1c -- offload/DeviceRTL/include/Allocator.h offload/DeviceRTL/include/Configuration.h offload/DeviceRTL/include/Debug.h offload/DeviceRTL/include/Interface.h offload/DeviceRTL/include/LibC.h offload/DeviceRTL/include/Mapping.h offload/DeviceRTL/include/State.h offload/DeviceRTL/include/Synchronization.h offload/DeviceRTL/include/Types.h offload/DeviceRTL/include/Utils.h offload/DeviceRTL/src/Allocator.cpp offload/DeviceRTL/src/Configuration.cpp offload/DeviceRTL/src/Debug.cpp offload/DeviceRTL/src/Kernel.cpp offload/DeviceRTL/src/LibC.cpp offload/DeviceRTL/src/Mapping.cpp offload/DeviceRTL/src/Misc.cpp offload/DeviceRTL/src/Parallelism.cpp offload/DeviceRTL/src/Reduction.cpp offload/DeviceRTL/src/State.cpp offload/DeviceRTL/src/Stub.cpp offload/DeviceRTL/src/Synchronization.cpp offload/DeviceRTL/src/Tasking.cpp offload/DeviceRTL/src/Utils.cpp offload/DeviceRTL/src/Workshare.cpp offload/include/DeviceImage.h offload/include/ExclusiveAccess.h offload/include/OffloadEntry.h offload/include/OffloadPolicy.h offload/include/OpenMP/InternalTypes.h offload/include/OpenMP/InteropAPI.h offload/include/OpenMP/Mapping.h offload/include/OpenMP/OMPT/Callback.h offload/include/OpenMP/OMPT/Connector.h offload/include/OpenMP/OMPT/Interface.h offload/include/OpenMP/omp.h offload/include/PluginManager.h offload/include/Shared/APITypes.h offload/include/Shared/Debug.h offload/include/Shared/Environment.h offload/include/Shared/EnvironmentVar.h offload/include/Shared/PluginAPI.h offload/include/Shared/PluginAPI.inc offload/include/Shared/Profile.h offload/include/Shared/Requirements.h offload/include/Shared/SourceInfo.h offload/include/Shared/Utils.h offload/include/Utils/ExponentialBackoff.h offload/include/device.h offload/include/omptarget.h offload/include/rtl.h offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa.cpp offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa.h offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h offload/plugins-nextgen/amdgpu/src/rtl.cpp offload/plugins-nextgen/amdgpu/utils/UtilitiesRTL.h offload/plugins-nextgen/common/OMPT/OmptCallback.cpp offload/plugins-nextgen/common/include/DLWrap.h offload/plugins-nextgen/common/include/GlobalHandler.h offload/plugins-nextgen/common/include/JIT.h offload/plugins-nextgen/common/include/MemoryManager.h offload/plugins-nextgen/common/include/PluginInterface.h offload/plugins-nextgen/common/include/RPC.h offload/plugins-nextgen/common/src/GlobalHandler.cpp offload/plugins-nextgen/common/src/JIT.cpp offload/plugins-nextgen/common/src/PluginInterface.cpp offload/plugins-nextgen/common/src/RPC.cpp offload/plugins-nextgen/common/src/Utils/ELF.cpp offload/plugins-nextgen/common/src/Utils/ELF.h offload/plugins-nextgen/cuda/dynamic_cuda/cuda.cpp offload/plugins-nextgen/cuda/dynamic_cuda/cuda.h offload/plugins-nextgen/cuda/src/rtl.cpp offload/plugins-nextgen/generic-elf-64bit/src/rtl.cpp offload/src/DeviceImage.cpp offload/src/LegacyAPI.cpp offload/src/OffloadRTL.cpp offload/src/OpenMP/API.cpp offload/src/OpenMP/InteropAPI.cpp offload/src/OpenMP/Mapping.cpp offload/src/OpenMP/OMPT/Callback.cpp offload/src/PluginManager.cpp offload/src/device.cpp offload/src/interface.cpp offload/src/omptarget.cpp offload/src/private.h offload/test/Inputs/declare_indirect_func.c offload/test/api/assert.c offload/test/api/is_initial_device.c offload/test/api/omp_device_managed_memory.c offload/test/api/omp_device_managed_memory_alloc.c offload/test/api/omp_device_memory.c offload/test/api/omp_dynamic_shared_memory.c offload/test/api/omp_dynamic_shared_memory_amdgpu.c offload/test/api/omp_dynamic_shared_memory_mixed.inc offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c offload/test/api/omp_dynamic_shared_memory_mixed_nvptx.c offload/test/api/omp_env_vars.c offload/test/api/omp_get_device_num.c offload/test/api/omp_get_mapped_ptr.c offload/test/api/omp_get_num_devices.c offload/test/api/omp_get_num_devices_with_empty_target.c offload/test/api/omp_get_num_procs.c offload/test/api/omp_host_pinned_memory.c offload/test/api/omp_host_pinned_memory_alloc.c offload/test/api/omp_indirect_call.c offload/test/api/omp_target_memcpy_async1.c offload/test/api/omp_target_memcpy_async2.c offload/test/api/omp_target_memcpy_rect_async1.c offload/test/api/omp_target_memcpy_rect_async2.c offload/test/api/omp_target_memset.c offload/test/api/ompx_3d.c offload/test/api/ompx_3d.cpp offload/test/api/ompx_sync.c offload/test/api/ompx_sync.cpp offload/test/env/base_ptr_ref_count.c offload/test/env/omp_target_debug.c offload/test/jit/empty_kernel.inc offload/test/jit/empty_kernel_lvl1.c offload/test/jit/empty_kernel_lvl2.c offload/test/jit/type_punning.c offload/test/libc/assert.c offload/test/libc/fwrite.c offload/test/libc/global_ctor_dtor.cpp offload/test/libc/host_call.c offload/test/libc/malloc.c offload/test/libc/puts.c offload/test/mapping/alloc_fail.c offload/test/mapping/array_section_implicit_capture.c offload/test/mapping/array_section_use_device_ptr.c offload/test/mapping/data_absent_at_exit.c offload/test/mapping/data_member_ref.cpp offload/test/mapping/declare_mapper_api.cpp offload/test/mapping/declare_mapper_nested_default_mappers.cpp offload/test/mapping/declare_mapper_nested_default_mappers_array.cpp offload/test/mapping/declare_mapper_nested_default_mappers_array_subscript.cpp offload/test/mapping/declare_mapper_nested_default_mappers_complex_structure.cpp offload/test/mapping/declare_mapper_nested_default_mappers_ptr_subscript.cpp offload/test/mapping/declare_mapper_nested_default_mappers_var.cpp offload/test/mapping/declare_mapper_nested_mappers.cpp offload/test/mapping/declare_mapper_target.cpp offload/test/mapping/declare_mapper_target_data.cpp offload/test/mapping/declare_mapper_target_data_enter_exit.cpp offload/test/mapping/declare_mapper_target_update.cpp offload/test/mapping/delete_inf_refcount.c offload/test/mapping/device_ptr_update.c offload/test/mapping/firstprivate_aligned.cpp offload/test/mapping/has_device_addr.cpp offload/test/mapping/implicit_device_ptr.c offload/test/mapping/is_device_ptr.cpp offload/test/mapping/lambda_by_value.cpp offload/test/mapping/lambda_mapping.cpp offload/test/mapping/low_alignment.c offload/test/mapping/map_back_race.cpp offload/test/mapping/ompx_hold/omp_target_disassociate_ptr.c offload/test/mapping/ompx_hold/struct.c offload/test/mapping/ompx_hold/target-data.c offload/test/mapping/ompx_hold/target.c offload/test/mapping/padding_not_mapped.c offload/test/mapping/power_of_two_alignment.c offload/test/mapping/pr38704.c offload/test/mapping/prelock.cpp offload/test/mapping/present/target.c offload/test/mapping/present/target_array_extension.c offload/test/mapping/present/target_data.c offload/test/mapping/present/target_data_array_extension.c offload/test/mapping/present/target_data_at_exit.c offload/test/mapping/present/target_enter_data.c offload/test/mapping/present/target_exit_data_delete.c offload/test/mapping/present/target_exit_data_release.c offload/test/mapping/present/target_update.c offload/test/mapping/present/target_update_array_extension.c offload/test/mapping/present/unified_shared_memory.c offload/test/mapping/present/zero_length_array_section.c offload/test/mapping/present/zero_length_array_section_exit.c offload/test/mapping/private_mapping.c offload/test/mapping/ptr_and_obj_motion.c offload/test/mapping/reduction_implicit_map.cpp offload/test/mapping/target_data_array_extension_at_exit.c offload/test/mapping/target_derefence_array_pointrs.cpp offload/test/mapping/target_has_device_addr.c offload/test/mapping/target_implicit_partial_map.c offload/test/mapping/target_map_for_member_data.cpp offload/test/mapping/target_pointers_members_map.cpp offload/test/mapping/target_update_array_extension.c offload/test/mapping/target_use_device_addr.c offload/test/mapping/target_uses_allocator.c offload/test/mapping/target_wrong_use_device_addr.c offload/test/offloading/assert.cpp offload/test/offloading/atomic-compare-signedness.c offload/test/offloading/back2back_distribute.c offload/test/offloading/barrier_fence.c offload/test/offloading/bug47654.cpp offload/test/offloading/bug49021.cpp offload/test/offloading/bug49334.cpp offload/test/offloading/bug49779.cpp offload/test/offloading/bug50022.cpp offload/test/offloading/bug51781.c offload/test/offloading/bug51982.c offload/test/offloading/bug53727.cpp offload/test/offloading/bug64959.c offload/test/offloading/bug64959_compile_only.c offload/test/offloading/bug74582.c offload/test/offloading/complex_reduction.cpp offload/test/offloading/ctor_dtor.cpp offload/test/offloading/cuda_no_devices.c offload/test/offloading/d2d_memcpy.c offload/test/offloading/default_thread_limit.c offload/test/offloading/dynamic_module.c offload/test/offloading/dynamic_module_load.c offload/test/offloading/extern.c offload/test/offloading/fortran/basic_array.c offload/test/offloading/generic_multiple_parallel_regions.c offload/test/offloading/global_constructor.cpp offload/test/offloading/host_as_target.c offload/test/offloading/indirect_fp_mapping.c offload/test/offloading/info.c offload/test/offloading/interop.c offload/test/offloading/lone_target_exit_data.c offload/test/offloading/looptripcnt.c offload/test/offloading/malloc.c offload/test/offloading/malloc_parallel.c offload/test/offloading/mandatory_but_no_devices.c offload/test/offloading/memory_manager.cpp offload/test/offloading/multiple_reductions_simple.c offload/test/offloading/non_contiguous_update.cpp offload/test/offloading/offloading_success.c offload/test/offloading/offloading_success.cpp offload/test/offloading/ompx_coords.c offload/test/offloading/ompx_saxpy_mixed.c offload/test/offloading/parallel_offloading_map.cpp offload/test/offloading/parallel_target_teams_reduction.cpp offload/test/offloading/parallel_target_teams_reduction_max.cpp offload/test/offloading/parallel_target_teams_reduction_min.cpp offload/test/offloading/requires.c offload/test/offloading/shared_lib_fp_mapping.c offload/test/offloading/small_trip_count.c offload/test/offloading/small_trip_count_thread_limit.cpp offload/test/offloading/spmdization.c offload/test/offloading/static_linking.c offload/test/offloading/std_complex_arithmetic.cpp offload/test/offloading/target-teams-atomic.c offload/test/offloading/target-tile.c offload/test/offloading/target_constexpr_mapping.cpp offload/test/offloading/target_critical_region.cpp offload/test/offloading/target_depend_nowait.cpp offload/test/offloading/target_map_for_member_data.cpp offload/test/offloading/target_nowait_target.cpp offload/test/offloading/task_in_reduction_target.c offload/test/offloading/taskloop_offload_nowait.cpp offload/test/offloading/test_libc.cpp offload/test/offloading/thread_limit.c offload/test/offloading/thread_state_1.c offload/test/offloading/thread_state_2.c offload/test/offloading/weak.c offload/test/offloading/wtime.c offload/test/ompt/callbacks.h offload/test/ompt/register_both.h offload/test/ompt/register_emi.h offload/test/ompt/register_emi_map.h offload/test/ompt/register_no_device_init.h offload/test/ompt/register_non_emi.h offload/test/ompt/register_non_emi_map.h offload/test/ompt/register_wrong_return.h offload/test/ompt/veccopy.c offload/test/ompt/veccopy_data.c offload/test/ompt/veccopy_disallow_both.c offload/test/ompt/veccopy_emi.c offload/test/ompt/veccopy_emi_map.c offload/test/ompt/veccopy_map.c offload/test/ompt/veccopy_no_device_init.c offload/test/ompt/veccopy_wrong_return.c offload/test/unified_shared_memory/api.c offload/test/unified_shared_memory/associate_ptr.c offload/test/unified_shared_memory/close_enter_exit.c offload/test/unified_shared_memory/close_manual.c offload/test/unified_shared_memory/close_member.c offload/test/unified_shared_memory/close_modifier.c offload/test/unified_shared_memory/shared_update.c offload/tools/deviceinfo/llvm-omp-device-info.cpp offload/tools/kernelreplay/llvm-omp-kernel-replay.cpp
``````````
</details>
<details>
<summary>
View the diff from clang-format here.
</summary>
``````````diff
diff --git a/offload/DeviceRTL/include/Types.h b/offload/DeviceRTL/include/Types.h
index 2e12d9da03..6d32f100c4 100644
--- a/offload/DeviceRTL/include/Types.h
+++ b/offload/DeviceRTL/include/Types.h
@@ -115,9 +115,9 @@ enum kmp_sched_t {
#define SCHEDULE_WITHOUT_MODIFIERS(s) \
(enum kmp_sched_t)( \
(s) & ~(kmp_sched_modifier_nonmonotonic | kmp_sched_modifier_monotonic))
-#define SCHEDULE_HAS_MONOTONIC(s) (((s)&kmp_sched_modifier_monotonic) != 0)
+#define SCHEDULE_HAS_MONOTONIC(s) (((s) & kmp_sched_modifier_monotonic) != 0)
#define SCHEDULE_HAS_NONMONOTONIC(s) \
- (((s)&kmp_sched_modifier_nonmonotonic) != 0)
+ (((s) & kmp_sched_modifier_nonmonotonic) != 0)
#define SCHEDULE_HAS_NO_MODIFIERS(s) \
(((s) & (kmp_sched_modifier_nonmonotonic | kmp_sched_modifier_monotonic)) == \
0)
diff --git a/offload/include/OpenMP/omp.h b/offload/include/OpenMP/omp.h
index b44c6aff1b..201ec5acaf 100644
--- a/offload/include/OpenMP/omp.h
+++ b/offload/include/OpenMP/omp.h
@@ -101,8 +101,9 @@ int __KAI_KMPC_CONVENTION omp_get_num_interop_properties(const omp_interop_t);
* The `omp_get_interop_int` routine retrieves an integer property from an
* `omp_interop_t` object.
*/
-omp_intptr_t __KAI_KMPC_CONVENTION
-omp_get_interop_int(const omp_interop_t, omp_interop_property_t, int *);
+omp_intptr_t __KAI_KMPC_CONVENTION omp_get_interop_int(const omp_interop_t,
+ omp_interop_property_t,
+ int *);
/*!
* The `omp_get_interop_ptr` routine retrieves a pointer property from an
* `omp_interop_t` object.
@@ -113,8 +114,9 @@ void *__KAI_KMPC_CONVENTION omp_get_interop_ptr(const omp_interop_t,
* The `omp_get_interop_str` routine retrieves a string property from an
* `omp_interop_t` object.
*/
-const char *__KAI_KMPC_CONVENTION
-omp_get_interop_str(const omp_interop_t, omp_interop_property_t, int *);
+const char *__KAI_KMPC_CONVENTION omp_get_interop_str(const omp_interop_t,
+ omp_interop_property_t,
+ int *);
/*!
* The `omp_get_interop_name` routine retrieves a property name from an
* `omp_interop_t` object.
diff --git a/offload/include/PluginManager.h b/offload/include/PluginManager.h
index a0499c3750..ecb0785e57 100644
--- a/offload/include/PluginManager.h
+++ b/offload/include/PluginManager.h
@@ -118,8 +118,10 @@ struct PluginManager {
// Unregister a shared library from all RTLs.
void unregisterLib(__tgt_bin_desc *Desc);
- void addDeviceImage(__tgt_bin_desc &TgtBinDesc, __tgt_device_image &TgtDeviceImage) {
- DeviceImages.emplace_back(std::make_unique<DeviceImageTy>(TgtBinDesc, TgtDeviceImage));
+ void addDeviceImage(__tgt_bin_desc &TgtBinDesc,
+ __tgt_device_image &TgtDeviceImage) {
+ DeviceImages.emplace_back(
+ std::make_unique<DeviceImageTy>(TgtBinDesc, TgtDeviceImage));
}
/// Return the device presented to the user as device \p DeviceNo if it is
diff --git a/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h b/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h
index 9c59d3bf82..4667eb7c85 100644
--- a/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h
+++ b/offload/plugins-nextgen/amdgpu/dynamic_hsa/hsa_ext_amd.h
@@ -103,11 +103,11 @@ hsa_status_t hsa_amd_agents_allow_access(uint32_t num_agents,
const uint32_t *flags,
const void *ptr);
-hsa_status_t hsa_amd_memory_lock(void* host_ptr, size_t size,
- hsa_agent_t* agents, int num_agent,
- void** agent_ptr);
+hsa_status_t hsa_amd_memory_lock(void *host_ptr, size_t size,
+ hsa_agent_t *agents, int num_agent,
+ void **agent_ptr);
-hsa_status_t hsa_amd_memory_unlock(void* host_ptr);
+hsa_status_t hsa_amd_memory_unlock(void *host_ptr);
hsa_status_t hsa_amd_memory_fill(void *ptr, uint32_t value, size_t count);
@@ -155,16 +155,15 @@ typedef enum {
typedef struct hsa_amd_pointer_info_s {
uint32_t size;
hsa_amd_pointer_type_t type;
- void* agentBaseAddress;
- void* hostBaseAddress;
+ void *agentBaseAddress;
+ void *hostBaseAddress;
size_t sizeInBytes;
} hsa_amd_pointer_info_t;
-hsa_status_t hsa_amd_pointer_info(const void* ptr,
- hsa_amd_pointer_info_t* info,
- void* (*alloc)(size_t),
- uint32_t* num_agents_accessible,
- hsa_agent_t** accessible);
+hsa_status_t hsa_amd_pointer_info(const void *ptr, hsa_amd_pointer_info_t *info,
+ void *(*alloc)(size_t),
+ uint32_t *num_agents_accessible,
+ hsa_agent_t **accessible);
#ifdef __cplusplus
}
diff --git a/offload/plugins-nextgen/common/include/DLWrap.h b/offload/plugins-nextgen/common/include/DLWrap.h
index 8934e7e701..00610678f9 100644
--- a/offload/plugins-nextgen/common/include/DLWrap.h
+++ b/offload/plugins-nextgen/common/include/DLWrap.h
@@ -107,7 +107,9 @@ template <size_t S> struct count {
static constexpr size_t N = count<S - 1>::N;
};
-template <> struct count<0> { static constexpr size_t N = 0; };
+template <> struct count<0> {
+ static constexpr size_t N = 0;
+};
// Get a constexpr size_t ID, starts at zero
#define DLWRAP_ID() (dlwrap::type::count<__LINE__>::N)
diff --git a/offload/test/api/omp_dynamic_shared_memory_amdgpu.c b/offload/test/api/omp_dynamic_shared_memory_amdgpu.c
index 0b4d9d6ea9..c4d7c8da2e 100644
--- a/offload/test/api/omp_dynamic_shared_memory_amdgpu.c
+++ b/offload/test/api/omp_dynamic_shared_memory_amdgpu.c
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm -openmp-opt-inline-device
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm
+// -openmp-opt-inline-device
// RUN: env LIBOMPTARGET_SHARED_MEMORY_SIZE=256 \
// RUN: %libomptarget-run-amdgcn-amd-amdhsa | %fcheck-amdgcn-amd-amdhsa
// REQUIRES: amdgcn-amd-amdhsa
diff --git a/offload/test/api/omp_dynamic_shared_memory_mixed.inc b/offload/test/api/omp_dynamic_shared_memory_mixed.inc
index 16e0becd09..08e050b935 100644
--- a/offload/test/api/omp_dynamic_shared_memory_mixed.inc
+++ b/offload/test/api/omp_dynamic_shared_memory_mixed.inc
@@ -9,8 +9,7 @@ int main() {
int Result[N], NumThreads;
#pragma omp target teams num_teams(1) thread_limit(N) \
- ompx_dyn_cgroup_mem(N * sizeof(Result[0])) \
- map(from : Result, NumThreads)
+ ompx_dyn_cgroup_mem(N * sizeof(Result[0])) map(from : Result, NumThreads)
{
int Buffer[N];
#pragma omp parallel
diff --git a/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c b/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c
index 656c3a20aa..9de28974d6 100644
--- a/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c
+++ b/offload/test/api/omp_dynamic_shared_memory_mixed_amdgpu.c
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm -openmp-opt-inline-device -I %S
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -O1 -mllvm
+// -openmp-opt-inline-device -I %S
// RUN: env LIBOMPTARGET_NEXTGEN_PLUGINS=1 \
// RUN: %libomptarget-run-amdgcn-amd-amdhsa | %fcheck-amdgcn-amd-amdhsa
// REQUIRES: amdgcn-amd-amdhsa
diff --git a/offload/test/api/omp_get_mapped_ptr.c b/offload/test/api/omp_get_mapped_ptr.c
index a8e11f912d..6d6fe42bc3 100644
--- a/offload/test/api/omp_get_mapped_ptr.c
+++ b/offload/test/api/omp_get_mapped_ptr.c
@@ -13,7 +13,7 @@ int main(int argc, char *argv[]) {
assert(device_ptr == NULL && "the pointer should not be mapped right now");
-#pragma omp target enter data map(to: host_data[:N])
+#pragma omp target enter data map(to : host_data[ : N])
device_ptr = omp_get_mapped_ptr(host_data, 0);
@@ -21,7 +21,7 @@ int main(int argc, char *argv[]) {
void *ptr = NULL;
-#pragma omp target map(from: ptr)
+#pragma omp target map(from : ptr)
{ ptr = host_data; }
assert(ptr == device_ptr && "wrong pointer mapping");
@@ -30,7 +30,7 @@ int main(int argc, char *argv[]) {
assert(device_ptr && "the pointer with offset should be mapped");
-#pragma omp target map(from: ptr)
+#pragma omp target map(from : ptr)
{ ptr = host_data + OFFSET; }
assert(ptr == device_ptr && "wrong pointer mapping");
diff --git a/offload/test/api/ompx_3d.c b/offload/test/api/ompx_3d.c
index 2ae17c226f..3045c1d2fc 100644
--- a/offload/test/api/ompx_3d.c
+++ b/offload/test/api/ompx_3d.c
@@ -6,8 +6,7 @@
void foo(int device) {
int tid = 0, bid = 0, bdim = 0;
-#pragma omp target teams distribute parallel for map(from \
- : tid, bid, bdim) \
+#pragma omp target teams distribute parallel for map(from : tid, bid, bdim) \
device(device) thread_limit(2) num_teams(5)
for (int i = 0; i < 1000; ++i) {
if (i == 42) {
diff --git a/offload/test/api/ompx_3d.cpp b/offload/test/api/ompx_3d.cpp
index 5b5491263a..f25b851532 100644
--- a/offload/test/api/ompx_3d.cpp
+++ b/offload/test/api/ompx_3d.cpp
@@ -6,8 +6,7 @@
void foo(int device) {
int tid = 0, bid = 0, bdim = 0;
-#pragma omp target teams distribute parallel for map(from \
- : tid, bid, bdim) \
+#pragma omp target teams distribute parallel for map(from : tid, bid, bdim) \
device(device) thread_limit(2) num_teams(5)
for (int i = 0; i < 1000; ++i) {
if (i == 42) {
diff --git a/offload/test/env/base_ptr_ref_count.c b/offload/test/env/base_ptr_ref_count.c
index 9a15568712..876991176d 100644
--- a/offload/test/env/base_ptr_ref_count.c
+++ b/offload/test/env/base_ptr_ref_count.c
@@ -1,5 +1,5 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic REQUIRES: libomptarget-debug
#include <stdio.h>
#include <stdlib.h>
diff --git a/offload/test/env/omp_target_debug.c b/offload/test/env/omp_target_debug.c
index ec81873a09..8694448acb 100644
--- a/offload/test/env/omp_target_debug.c
+++ b/offload/test/env/omp_target_debug.c
@@ -1,6 +1,8 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=DEBUG
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=0 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=NDEBUG
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty
+// -check-prefix=DEBUG RUN: %libomptarget-compile-generic && env
+// LIBOMPTARGET_DEBUG=0 %libomptarget-run-generic 2>&1 | %fcheck-generic
+// -allow-empty -check-prefix=NDEBUG REQUIRES: libomptarget-debug
int main(void) {
#pragma omp target
diff --git a/offload/test/jit/empty_kernel.inc b/offload/test/jit/empty_kernel.inc
index 43813891cc..3e60ca994a 100644
--- a/offload/test/jit/empty_kernel.inc
+++ b/offload/test/jit/empty_kernel.inc
@@ -1,15 +1,14 @@
-int main(int argc, char** argv) {
- #pragma omp TGT1_DIRECTIVE
+int main(int argc, char **argv) {
+#pragma omp TGT1_DIRECTIVE
{
#ifdef LOOP_DIRECTIVE
- #pragma omp LOOP_DIRECTIVE
+#pragma omp LOOP_DIRECTIVE
for (int i = 0; i < argc; ++i)
#endif
{
#ifdef BODY_DIRECTIVE
- #pragma omp BODY_DIRECTIVE
- {
- }
+#pragma omp BODY_DIRECTIVE
+ {}
#endif
}
}
@@ -18,14 +17,13 @@ int main(int argc, char** argv) {
#pragma omp TGT2_DIRECTIVE
{
#ifdef LOOP_DIRECTIVE
- #pragma omp LOOP_DIRECTIVE
+#pragma omp LOOP_DIRECTIVE
for (int i = 0; i < argc; ++i)
#endif
{
#ifdef BODY_DIRECTIVE
- #pragma omp BODY_DIRECTIVE
- {
- }
+#pragma omp BODY_DIRECTIVE
+ {}
#endif
}
}
diff --git a/offload/test/mapping/alloc_fail.c b/offload/test/mapping/alloc_fail.c
index c4ae70fc73..31912d4bdb 100644
--- a/offload/test/mapping/alloc_fail.c
+++ b/offload/test/mapping/alloc_fail.c
@@ -2,9 +2,11 @@
// RUN: %libomptarget-run-fail-generic 2>&1 \
// RUN: | %fcheck-generic
-// CHECK: omptarget message: explicit extension not allowed: host address specified is 0x{{.*}} (8 bytes), but device allocation maps to host at 0x{{.*}} (8 bytes)
-// CHECK: omptarget error: Call to getTargetPointer returned null pointer (device failure or illegal mapping).
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget message: explicit extension not allowed: host address
+// specified is 0x{{.*}} (8 bytes), but device allocation maps to host at
+// 0x{{.*}} (8 bytes) CHECK: omptarget error: Call to getTargetPointer returned
+// null pointer (device failure or illegal mapping). CHECK: omptarget fatal
+// error 1: failure of target construct while offloading is mandatory
int main() {
int arr[4] = {0, 1, 2, 3};
diff --git a/offload/test/mapping/array_section_implicit_capture.c b/offload/test/mapping/array_section_implicit_capture.c
index 210b7e51cb..1042bd23bd 100644
--- a/offload/test/mapping/array_section_implicit_capture.c
+++ b/offload/test/mapping/array_section_implicit_capture.c
@@ -1,4 +1,4 @@
-// RUN: %libomptarget-compile-generic
+// RUN: %libomptarget-compile-generic
// RUN: %libomptarget-run-generic 2>&1 \
// RUN: | %fcheck-generic
diff --git a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
index c6c5657ae6..6408dae555 100644
--- a/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_default_mappers.cpp
@@ -44,8 +44,8 @@ int main() {
int spp00fa = -1, spp00fca = -1, spp00fb_r = -1;
__intptr_t p = reinterpret_cast<__intptr_t>(&x[0]);
-#pragma omp target map(tofrom: spp[0][0]) firstprivate(p) \
- map(from: spp00fa, spp00fca, spp00fb_r)
+#pragma omp target map(tofrom : spp[0][0]) firstprivate(p) \
+ map(from : spp00fa, spp00fca, spp00fb_r)
{
spp00fa = spp[0][0].f.a;
spp00fca = spp[0][0].f.c.a;
diff --git a/offload/test/mapping/declare_mapper_nested_mappers.cpp b/offload/test/mapping/declare_mapper_nested_mappers.cpp
index a9e3f05e0f..cbcb10bbf1 100644
--- a/offload/test/mapping/declare_mapper_nested_mappers.cpp
+++ b/offload/test/mapping/declare_mapper_nested_mappers.cpp
@@ -42,8 +42,8 @@ int main() {
int spp00fa = -1, spp00fb_r = -1, spp00fg1 = -1, spp00fg_r = -1;
__intptr_t p = reinterpret_cast<__intptr_t>(&x[0]),
p1 = reinterpret_cast<__intptr_t>(&y[0]);
-#pragma omp target map(tofrom : spp[0][0]) firstprivate(p, p1) \
- map(from: spp00fa, spp00fb_r, spp00fg1, spp00fg_r)
+#pragma omp target map(tofrom : spp[0][0]) firstprivate(p, p1) \
+ map(from : spp00fa, spp00fb_r, spp00fg1, spp00fg_r)
{
spp00fa = spp[0][0].f.a;
spp00fb_r = spp[0][0].f.b == reinterpret_cast<void *>(p) ? 1 : 0;
@@ -56,7 +56,7 @@ int main() {
spp[0][0].f.b[1] = 40;
spp[0][0].g[1] = 50;
}
- printf("%d %d %d %d\n", spp00fa, spp00fb_r, spp00fg1, spp00fg_r);
+ printf("%d %d %d %d\n", spp00fa, spp00fb_r, spp00fg1, spp00fg_r);
// CHECK: 222 0 30 0
printf("%d %d %4.5f %d %d %d\n", spp[0][0].e, spp[0][0].f.a, spp[0][0].f.b[1],
spp[0][0].f.b == &x[0] ? 1 : 0, spp[0][0].g[1],
diff --git a/offload/test/mapping/device_ptr_update.c b/offload/test/mapping/device_ptr_update.c
index c6719d2285..77aab6c579 100644
--- a/offload/test/mapping/device_ptr_update.c
+++ b/offload/test/mapping/device_ptr_update.c
@@ -15,7 +15,8 @@ int main(void) {
s1.p = A;
-// DEBUG: Update pointer ([[DEV_PTR:0x[^ ]+]]) -> {{\[}}[[DEV_OBJ_A:0x[^ ]+]]{{\]}}
+// DEBUG: Update pointer ([[DEV_PTR:0x[^ ]+]]) -> {{\[}}[[DEV_OBJ_A:0x[^
+// ]+]]{{\]}}
#pragma omp target enter data map(alloc : s1.p[0 : 10])
// DEBUG-NOT: Update pointer ([[DEV_PTR]]) -> {{\[}}[[DEV_OBJ_A]]{{\]}}
diff --git a/offload/test/mapping/padding_not_mapped.c b/offload/test/mapping/padding_not_mapped.c
index 3ee70ab640..99f57f4fbb 100644
--- a/offload/test/mapping/padding_not_mapped.c
+++ b/offload/test/mapping/padding_not_mapped.c
@@ -18,7 +18,11 @@
#include <stdio.h>
int main() {
- struct S { int x; int y; double z; } s = {1, 2, 3};
+ struct S {
+ int x;
+ int y;
+ double z;
+ } s = {1, 2, 3};
// CHECK: &s.x = 0x[[#%x,HOST_ADDR:]], size = [[#%u,SIZE:]]
fprintf(stderr, "&s = %p\n", &s);
@@ -26,18 +30,20 @@ int main() {
fprintf(stderr, "&s.y = %p\n", &s.y);
fprintf(stderr, "&s.z = %p\n", &s.z);
- // CHECK: s.x is present: 0
- // CHECK: s.x = 1{{$}}
- #pragma omp target enter data map(alloc: s.y, s.z)
+// CHECK: s.x is present: 0
+// CHECK: s.x = 1{{$}}
+#pragma omp target enter data map(alloc : s.y, s.z)
int dev = omp_get_default_device();
fprintf(stderr, "s.x is present: %d\n", omp_target_is_present(&s.x, dev));
- #pragma omp target update from(s.x) // should have no effect
+#pragma omp target update from(s.x) // should have no effect
fprintf(stderr, "s.x = %d\n", s.x);
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
- // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
- #pragma omp target enter data map(present, alloc: s.x)
+// CHECK: omptarget message: device mapping required by 'present' map type
+// modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+// bytes) CHECK: omptarget error: Call to getTargetPointer returned null pointer
+// ('present' map type modifier). CHECK: omptarget fatal error 1: failure of
+// target construct while offloading is mandatory
+#pragma omp target enter data map(present, alloc : s.x)
return 0;
}
diff --git a/offload/test/mapping/power_of_two_alignment.c b/offload/test/mapping/power_of_two_alignment.c
index faebe4f89f..b31c43ac51 100644
--- a/offload/test/mapping/power_of_two_alignment.c
+++ b/offload/test/mapping/power_of_two_alignment.c
@@ -42,10 +42,13 @@
#include <stdint.h>
#include <stdio.h>
-template <typename StackPad>
-void test() {
+template <typename StackPad> void test() {
StackPad stackPad;
- struct S { char x; char y[7]; char z[8]; };
+ struct S {
+ char x;
+ char y[7];
+ char z[8];
+ };
struct S collidePre, s, collidePost;
uintptr_t mod16 = (uintptr_t)&s % 16;
fprintf(stderr, "&s = %p\n", &s);
@@ -56,8 +59,8 @@ void test() {
}
fprintf(stderr, "&collidePre = %p\n", &collidePre);
fprintf(stderr, "&collidePost = %p\n", &collidePost);
- #pragma omp target data map(to:s.y, s.z)
- #pragma omp target data map(to:collidePre, collidePost)
+#pragma omp target data map(to : s.y, s.z)
+#pragma omp target data map(to : collidePre, collidePost)
;
}
diff --git a/offload/test/mapping/present/target.c b/offload/test/mapping/present/target.c
index 4344c42c80..e5cab61374 100644
--- a/offload/test/mapping/present/target.c
+++ b/offload/test/mapping/present/target.c
@@ -18,11 +18,13 @@ int main() {
// CHECK: i is present
fprintf(stderr, "i is present\n");
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
- // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
- // CHECK: omptarget error: Call to targetDataBegin failed, abort target.
- // CHECK: omptarget error: Failed to process data before launching the kernel.
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: device mapping required by 'present' map type
+ // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+ // bytes) CHECK: omptarget error: Call to getTargetPointer returned null
+ // pointer ('present' map type modifier). CHECK: omptarget error: Call to
+ // targetDataBegin failed, abort target. CHECK: omptarget error: Failed to
+ // process data before launching the kernel. CHECK: omptarget fatal error 1:
+ // failure of target construct while offloading is mandatory
#pragma omp target map(present, alloc : i)
;
diff --git a/offload/test/mapping/present/target_array_extension.c b/offload/test/mapping/present/target_array_extension.c
index 873b2b3617..3e3d991b40 100644
--- a/offload/test/mapping/present/target_array_extension.c
+++ b/offload/test/mapping/present/target_array_extension.c
@@ -70,12 +70,16 @@ int main() {
// CHECK: arr is present
fprintf(stderr, "arr is present\n");
- // CHECK: omptarget message: explicit extension not allowed: host address specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes)
- // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
- // CHECK: omptarget error: Call to targetDataBegin failed, abort target.
- // CHECK: omptarget error: Failed to process data before launching the kernel.
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: explicit extension not allowed: host address
+ // specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device
+ // allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
+ // CHECK: omptarget message: device mapping required by 'present' map type
+ // modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]]
+ // ([[#LARGE_BYTES]] bytes) CHECK: omptarget error: Call to getTargetPointer
+ // returned null pointer ('present' map type modifier). CHECK: omptarget
+ // error: Call to targetDataBegin failed, abort target. CHECK: omptarget
+ // error: Failed to process data before launching the kernel. CHECK: omptarget
+ // fatal error 1: failure of target construct while offloading is mandatory
#pragma omp target data map(alloc : arr[SMALL])
{
#pragma omp target map(present, tofrom : arr[LARGE])
diff --git a/offload/test/mapping/present/target_data.c b/offload/test/mapping/present/target_data.c
index f894283a0c..ae3b762c34 100644
--- a/offload/test/mapping/present/target_data.c
+++ b/offload/test/mapping/present/target_data.c
@@ -18,8 +18,10 @@ int main() {
// CHECK: i is present
fprintf(stderr, "i is present\n");
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: device mapping required by 'present' map type
+ // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+ // bytes) CHECK: omptarget fatal error 1: failure of target construct while
+ // offloading is mandatory
#pragma omp target data map(present, alloc : i)
;
diff --git a/offload/test/mapping/present/target_data_array_extension.c b/offload/test/mapping/present/target_data_array_extension.c
index 794543a246..828351ef6b 100644
--- a/offload/test/mapping/present/target_data_array_extension.c
+++ b/offload/test/mapping/present/target_data_array_extension.c
@@ -70,10 +70,14 @@ int main() {
// CHECK: arr is present
fprintf(stderr, "arr is present\n");
- // CHECK: omptarget message: explicit extension not allowed: host address specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes)
- // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: explicit extension not allowed: host address
+ // specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device
+ // allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
+ // CHECK: omptarget message: device mapping required by 'present' map type
+ // modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]]
+ // ([[#LARGE_BYTES]] bytes) CHECK: omptarget error: Call to getTargetPointer
+ // returned null pointer ('present' map type modifier). CHECK: omptarget fatal
+ // error 1: failure of target construct while offloading is mandatory
#pragma omp target data map(alloc : arr[SMALL])
{
#pragma omp target data map(present, tofrom : arr[LARGE])
diff --git a/offload/test/mapping/present/target_enter_data.c b/offload/test/mapping/present/target_enter_data.c
index 871a05290e..633277ddf0 100644
--- a/offload/test/mapping/present/target_enter_data.c
+++ b/offload/test/mapping/present/target_enter_data.c
@@ -18,9 +18,11 @@ int main() {
// CHECK: i is present
fprintf(stderr, "i is present\n");
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
- // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: device mapping required by 'present' map type
+ // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+ // bytes) CHECK: omptarget error: Call to getTargetPointer returned null
+ // pointer ('present' map type modifier). CHECK: omptarget fatal error 1:
+ // failure of target construct while offloading is mandatory
#pragma omp target enter data map(present, alloc : i)
// CHECK-NOT: i is present
diff --git a/offload/test/mapping/present/target_exit_data_delete.c b/offload/test/mapping/present/target_exit_data_delete.c
index 0fb812b299..d791166147 100644
--- a/offload/test/mapping/present/target_exit_data_delete.c
+++ b/offload/test/mapping/present/target_exit_data_delete.c
@@ -17,8 +17,10 @@ int main() {
// CHECK: i was present
fprintf(stderr, "i was present\n");
-// CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget message: device mapping required by 'present' map type
+// modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+// bytes) CHECK: omptarget fatal error 1: failure of target construct while
+// offloading is mandatory
#pragma omp target exit data map(present, delete : i)
// CHECK-NOT: i was present
diff --git a/offload/test/mapping/present/target_exit_data_release.c b/offload/test/mapping/present/target_exit_data_release.c
index 14be22faba..083f6c080d 100644
--- a/offload/test/mapping/present/target_exit_data_release.c
+++ b/offload/test/mapping/present/target_exit_data_release.c
@@ -17,8 +17,10 @@ int main() {
// CHECK: i was present
fprintf(stderr, "i was present\n");
-// CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget message: device mapping required by 'present' map type
+// modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+// bytes) CHECK: omptarget fatal error 1: failure of target construct while
+// offloading is mandatory
#pragma omp target exit data map(present, release : i)
// CHECK-NOT: i was present
diff --git a/offload/test/mapping/present/target_update.c b/offload/test/mapping/present/target_update.c
index 9f6783b6ef..5df4eee6fc 100644
--- a/offload/test/mapping/present/target_update.c
+++ b/offload/test/mapping/present/target_update.c
@@ -32,8 +32,10 @@ int main() {
// CHECK: i is present
fprintf(stderr, "i is present\n");
- // CHECK: omptarget message: device mapping required by 'present' motion modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: device mapping required by 'present' motion
+ // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+ // bytes) CHECK: omptarget fatal error 1: failure of target construct while
+ // offloading is mandatory
#pragma omp target update CLAUSE(present : i)
// CHECK-NOT: i is present
diff --git a/offload/test/mapping/present/target_update_array_extension.c b/offload/test/mapping/present/target_update_array_extension.c
index 11ad4a8d49..f8f3c94d21 100644
--- a/offload/test/mapping/present/target_update_array_extension.c
+++ b/offload/test/mapping/present/target_update_array_extension.c
@@ -66,8 +66,10 @@ int main() {
// CHECK: arr is present
fprintf(stderr, "arr is present\n");
- // CHECK: omptarget message: device mapping required by 'present' motion modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes)
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: device mapping required by 'present' motion
+ // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]]
+ // bytes) CHECK: omptarget fatal error 1: failure of target construct while
+ // offloading is mandatory
#pragma omp target data map(alloc : arr[SMALL])
{
#pragma omp target update CLAUSE(present : arr[LARGE])
diff --git a/offload/test/mapping/present/zero_length_array_section.c b/offload/test/mapping/present/zero_length_array_section.c
index e903a268ef..4759585845 100644
--- a/offload/test/mapping/present/zero_length_array_section.c
+++ b/offload/test/mapping/present/zero_length_array_section.c
@@ -20,11 +20,13 @@ int main() {
// arr[0:0] doesn't create an actual mapping in the first directive.
//
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
- // CHECK: omptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
- // CHECK: omptarget error: Call to targetDataBegin failed, abort target.
- // CHECK: omptarget error: Failed to process data before launching the kernel.
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: device mapping required by 'present' map type
+ // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
+ // CHECK: omptarget error: Call to getTargetPointer returned null pointer
+ // ('present' map type modifier). CHECK: omptarget error: Call to
+ // targetDataBegin failed, abort target. CHECK: omptarget error: Failed to
+ // process data before launching the kernel. CHECK: omptarget fatal error 1:
+ // failure of target construct while offloading is mandatory
#pragma omp target data map(alloc : arr[0 : 0])
#pragma omp target map(present, alloc : arr[0 : 0])
;
diff --git a/offload/test/mapping/present/zero_length_array_section_exit.c b/offload/test/mapping/present/zero_length_array_section_exit.c
index 5a7360542e..c88c47f115 100644
--- a/offload/test/mapping/present/zero_length_array_section_exit.c
+++ b/offload/test/mapping/present/zero_length_array_section_exit.c
@@ -19,8 +19,10 @@ int main() {
// arr[0:0] doesn't create an actual mapping in the first directive.
//
- // CHECK: omptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
- // CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+ // CHECK: omptarget message: device mapping required by 'present' map type
+ // modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes)
+ // CHECK: omptarget fatal error 1: failure of target construct while
+ // offloading is mandatory
#pragma omp target enter data map(alloc : arr[0 : 0])
#pragma omp target exit data map(present, release : arr[0 : 0])
diff --git a/offload/test/mapping/private_mapping.c b/offload/test/mapping/private_mapping.c
index 1329a66a5d..157cc4c511 100644
--- a/offload/test/mapping/private_mapping.c
+++ b/offload/test/mapping/private_mapping.c
@@ -10,9 +10,10 @@ int main() {
int data3[3] = {100, 200, 500};
int sum[16] = {0};
- for (int i=0; i<16; i++) sum[i] = 10000;
+ for (int i = 0; i < 16; i++)
+ sum[i] = 10000;
-#pragma omp target teams distribute parallel for map(tofrom : sum[:16]) \
+#pragma omp target teams distribute parallel for map(tofrom : sum[ : 16]) \
firstprivate(data1, data2, data3)
for (int i = 0; i < 16; ++i) {
for (int j = 0; j < 3; ++j) {
diff --git a/offload/test/mapping/target_wrong_use_device_addr.c b/offload/test/mapping/target_wrong_use_device_addr.c
index 6c7939ea19..13415d97c7 100644
--- a/offload/test/mapping/target_wrong_use_device_addr.c
+++ b/offload/test/mapping/target_wrong_use_device_addr.c
@@ -11,7 +11,7 @@ int main() {
// CHECK: host addr=0x[[#%x,HOST_ADDR:]]
fprintf(stderr, "host addr=%p\n", x);
-#pragma omp target data map(to : x [0:10])
+#pragma omp target data map(to : x[0 : 10])
{
// CHECK: omptarget device 0 info: variable x does not have a valid device
// counterpart
@@ -24,4 +24,3 @@ int main() {
return 0;
}
-
diff --git a/offload/test/offloading/back2back_distribute.c b/offload/test/offloading/back2back_distribute.c
index 63cabd0c66..51fd844b32 100644
--- a/offload/test/offloading/back2back_distribute.c
+++ b/offload/test/offloading/back2back_distribute.c
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compile-generic -O3 && %libomptarget-run-generic | %fcheck-generic
+// RUN: %libomptarget-compile-generic -O3 && %libomptarget-run-generic |
+// %fcheck-generic
#include <omp.h>
#include <stdio.h>
@@ -7,10 +8,10 @@
#define MAX_N 25000
void reset_input(double *a, double *a_h, double *b, double *c) {
- for(int i = 0 ; i < MAX_N ; i++) {
+ for (int i = 0; i < MAX_N; i++) {
a[i] = a_h[i] = i;
- b[i] = i*2;
- c[i] = i-3;
+ b[i] = i * 2;
+ c[i] = i - 3;
}
}
@@ -22,15 +23,16 @@ int main(int argc, char *argv[]) {
double *b = (double *)calloc(MAX_N, sizeof(double));
double *c = (double *)calloc(MAX_N, sizeof(double));
-#pragma omp target enter data map(to:a[:MAX_N],b[:MAX_N],c[:MAX_N],d[:MAX_N])
+#pragma omp target enter data map(to : a[ : MAX_N], b[ : MAX_N], c[ : MAX_N], \
+ d[ : MAX_N])
- for (int n = 32 ; n < MAX_N ; n+=5000) {
+ for (int n = 32; n < MAX_N; n += 5000) {
reset_input(a, a_h, b, c);
-#pragma omp target update to(a[:n],b[:n],c[:n],d[:n])
+#pragma omp target update to(a[ : n], b[ : n], c[ : n], d[ : n])
int t = 0;
- for (int tms = 1 ; tms <= 256 ; tms *= 2) { // 8 times
- for (int ths = 32 ; ths <= 1024 ; ths *= 2) { // 6 times
+ for (int tms = 1; tms <= 256; tms *= 2) { // 8 times
+ for (int ths = 32; ths <= 1024; ths *= 2) { // 6 times
t++;
#pragma omp target
#pragma omp teams num_teams(tms) thread_limit(ths)
@@ -41,29 +43,31 @@ int main(int argc, char *argv[]) {
}
#pragma omp distribute parallel for
for (int i = 0; i < n; ++i) {
- d[i] -= b[i] + c[i];
+ d[i] -= b[i] + c[i];
}
}
} // loop over 'ths'
- } // loop over 'tms'
+ } // loop over 'tms'
// check results for each 'n'
- for (int times = 0 ; times < t ; times++) {
+ for (int times = 0; times < t; times++) {
for (int i = 0; i < n; ++i) {
a_h[i] += b[i] + c[i];
}
for (int i = 0; i < n; ++i)
d_h[i] -= b[i] + c[i];
}
-#pragma omp target update from(a[:n],d[:n])
+#pragma omp target update from(a[ : n], d[ : n])
for (int i = 0; i < n; ++i) {
if (a_h[i] != a[i]) {
- printf("A Error at n = %d, i = %d: host = %f, device = %f\n", n, i, a_h[i], a[i]);
+ printf("A Error at n = %d, i = %d: host = %f, device = %f\n", n, i,
+ a_h[i], a[i]);
return 1;
}
if (d_h[i] != d[i]) {
- printf("D Error at n = %d, i = %d: host = %lf, device = %lf\n", n, i, d_h[i], d[i]);
+ printf("D Error at n = %d, i = %d: host = %lf, device = %lf\n", n, i,
+ d_h[i], d[i]);
return 1;
}
}
diff --git a/offload/test/offloading/bug49021.cpp b/offload/test/offloading/bug49021.cpp
index 37da8ce5d8..8b03c4580e 100644
--- a/offload/test/offloading/bug49021.cpp
+++ b/offload/test/offloading/bug49021.cpp
@@ -1,7 +1,8 @@
// RUN: %libomptarget-compilexx-generic -O3 && %libomptarget-run-generic
-// RUN: %libomptarget-compilexx-generic -O3 -ffast-math && %libomptarget-run-generic
-// RUN: %libomptarget-compileoptxx-generic -O3 && %libomptarget-run-generic
-// RUN: %libomptarget-compileoptxx-generic -O3 -ffast-math && %libomptarget-run-generic
+// RUN: %libomptarget-compilexx-generic -O3 -ffast-math &&
+// %libomptarget-run-generic RUN: %libomptarget-compileoptxx-generic -O3 &&
+// %libomptarget-run-generic RUN: %libomptarget-compileoptxx-generic -O3
+// -ffast-math && %libomptarget-run-generic
#include <iostream>
diff --git a/offload/test/offloading/bug49334.cpp b/offload/test/offloading/bug49334.cpp
index a22d3fe9f6..f5f2ade80a 100644
--- a/offload/test/offloading/bug49334.cpp
+++ b/offload/test/offloading/bug49334.cpp
@@ -91,7 +91,7 @@ int BlockMatMul_TargetNowait(BlockMatrix &A, BlockMatrix &B, BlockMatrix &C) {
for (int k = 0; k < N / BS; ++k) {
float *BlockA = A.GetBlock(i, k);
float *BlockB = B.GetBlock(k, j);
-// clang-format off
+ // clang-format off
#pragma omp target depend(in: BlockA[0], BlockB[0]) depend(inout: BlockC[0]) \
map(to: BlockA[:BS * BS], BlockB[:BS * BS]) \
map(tofrom: BlockC[:BS * BS]) nowait
diff --git a/offload/test/offloading/complex_reduction.cpp b/offload/test/offloading/complex_reduction.cpp
index 3239f48743..32f06ae4cb 100644
--- a/offload/test/offloading/complex_reduction.cpp
+++ b/offload/test/offloading/complex_reduction.cpp
@@ -1,5 +1,6 @@
// RUN: %libomptarget-compilexx-generic -O3 && %libomptarget-run-generic
-// RUN: %libomptarget-compilexx-generic -O3 -ffast-math && %libomptarget-run-generic
+// RUN: %libomptarget-compilexx-generic -O3 -ffast-math &&
+// %libomptarget-run-generic
#include <complex>
#include <iostream>
@@ -20,8 +21,8 @@ template <typename T> void test_map() {
}
#if !defined(__NO_UDR)
-#pragma omp declare reduction(+ : std::complex <float> : omp_out += omp_in)
-#pragma omp declare reduction(+ : std::complex <double> : omp_out += omp_in)
+#pragma omp declare reduction(+ : std::complex<float> : omp_out += omp_in)
+#pragma omp declare reduction(+ : std::complex<double> : omp_out += omp_in)
#endif
template <typename T> class initiator {
@@ -43,7 +44,8 @@ template <typename T> void test_reduction() {
sum_host += array[i];
}
-#pragma omp target teams distribute parallel for map(to : array[:size]) reduction(+ : sum)
+#pragma omp target teams distribute parallel for map(to : array[ : size]) \
+ reduction(+ : sum)
for (int i = 0; i < size; i++)
sum += array[i];
@@ -55,10 +57,8 @@ template <typename T> void test_reduction() {
const int nblock(10), block_size(10);
T block_sum[nblock];
-#pragma omp target teams distribute map(to \
- : array[:size]) \
- map(from \
- : block_sum[:nblock])
+#pragma omp target teams distribute map(to : array[ : size]) \
+ map(from : block_sum[ : nblock])
for (int ib = 0; ib < nblock; ib++) {
T partial_sum(0);
const int istart = ib * block_size;
diff --git a/offload/test/offloading/dynamic_module_load.c b/offload/test/offloading/dynamic_module_load.c
index 935d402ef2..44728200bd 100644
--- a/offload/test/offloading/dynamic_module_load.c
+++ b/offload/test/offloading/dynamic_module_load.c
@@ -1,4 +1,6 @@
-// RUN: %libomptarget-compile-generic -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t -ldl && %libomptarget-run-generic %t.so 2>&1 | %fcheck-generic
+// RUN: %libomptarget-compile-generic -DSHARED -fPIC -shared -o %t.so && %clang
+// %flags %s -o %t -ldl && %libomptarget-run-generic %t.so 2>&1 |
+// %fcheck-generic
#ifdef SHARED
#include <stdio.h>
diff --git a/offload/test/offloading/extern.c b/offload/test/offloading/extern.c
index 35cf905ee0..df9285ce61 100644
--- a/offload/test/offloading/extern.c
+++ b/offload/test/offloading/extern.c
@@ -1,11 +1,12 @@
// RUN: %libomptarget-compile-generic -DVAR -c -o %t.o
-// RUN: %libomptarget-compile-generic %t.o && %libomptarget-run-generic | %fcheck-generic
+// RUN: %libomptarget-compile-generic %t.o && %libomptarget-run-generic |
+// %fcheck-generic
#ifdef VAR
int x = 1;
#else
-#include <stdio.h>
#include <assert.h>
+#include <stdio.h>
extern int x;
int main() {
@@ -22,6 +23,6 @@ int main() {
assert(value == 999);
// CHECK: PASS
- printf ("PASS\n");
+ printf("PASS\n");
}
#endif
diff --git a/offload/test/offloading/fortran/basic_array.c b/offload/test/offloading/fortran/basic_array.c
index 4daf1d4b0a..0928814bdd 100644
--- a/offload/test/offloading/fortran/basic_array.c
+++ b/offload/test/offloading/fortran/basic_array.c
@@ -14,7 +14,7 @@ void increment_at(int i, int *array);
#pragma omp end declare target
void increment_array(int *b, int n) {
-#pragma omp target map(tofrom : b [0:n])
+#pragma omp target map(tofrom : b[0 : n])
for (int i = 0; i < n; i++) {
increment_at(i, b);
}
diff --git a/offload/test/offloading/global_constructor.cpp b/offload/test/offloading/global_constructor.cpp
index eb68c5f783..6260d8994c 100644
--- a/offload/test/offloading/global_constructor.cpp
+++ b/offload/test/offloading/global_constructor.cpp
@@ -1,4 +1,5 @@
-// RUN: %libomptarget-compilexx-generic && %libomptarget-run-generic | %fcheck-generic
+// RUN: %libomptarget-compilexx-generic && %libomptarget-run-generic |
+// %fcheck-generic
#include <cstdio>
diff --git a/offload/test/offloading/info.c b/offload/test/offloading/info.c
index 81300cc92e..89efadc547 100644
--- a/offload/test/offloading/info.c
+++ b/offload/test/offloading/info.c
@@ -22,7 +22,7 @@ int main() {
int C[N];
int val = 1;
-// clang-format off
+ // clang-format off
// INFO: info: Entering OpenMP data region with being_mapper at info.c:{{[0-9]+}}:{{[0-9]+}} with 3 arguments:
// INFO: info: alloc(A[0:64])[256]
// INFO: info: tofrom(B[0:64])[256]
@@ -57,7 +57,7 @@ int main() {
// INFO: info: OpenMP Host-Device pointer mappings after block at info.c:[[#%u,]]:[[#%u,]]:
// INFO: info: Host Ptr Target Ptr Size (B) DynRefCount HoldRefCount Declaration
// INFO: info: [[#%#x,]] [[#%#x,]] 4 INF 0 global at unknown:0:0
-// clang-format on
+ // clang-format on
#pragma omp target data map(alloc : A[0 : N]) \
map(ompx_hold, tofrom : B[0 : N]) map(to : C[0 : N])
#pragma omp target firstprivate(val)
diff --git a/offload/test/offloading/looptripcnt.c b/offload/test/offloading/looptripcnt.c
index d1a095038b..5547bbac16 100644
--- a/offload/test/offloading/looptripcnt.c
+++ b/offload/test/offloading/looptripcnt.c
@@ -1,5 +1,6 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=DEBUG
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty
+// -check-prefix=DEBUG REQUIRES: libomptarget-debug
/*
Test for looptripcount being popped from runtime stack.
diff --git a/offload/test/offloading/mandatory_but_no_devices.c b/offload/test/offloading/mandatory_but_no_devices.c
index ecdee72aca..c44b3f6ba8 100644
--- a/offload/test/offloading/mandatory_but_no_devices.c
+++ b/offload/test/offloading/mandatory_but_no_devices.c
@@ -47,7 +47,8 @@
#include <omp.h>
#include <stdio.h>
-// CHECK: omptarget fatal error 1: failure of target construct while offloading is mandatory
+// CHECK: omptarget fatal error 1: failure of target construct while offloading
+// is mandatory
int main(void) {
int X;
#pragma omp DIR device(omp_get_initial_device())
diff --git a/offload/test/offloading/non_contiguous_update.cpp b/offload/test/offloading/non_contiguous_update.cpp
index 609f0f967f..3d9ce82361 100644
--- a/offload/test/offloading/non_contiguous_update.cpp
+++ b/offload/test/offloading/non_contiguous_update.cpp
@@ -1,5 +1,6 @@
-// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=DEBUG
-// REQUIRES: libomptarget-debug
+// RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1
+// %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty
+// -check-prefix=DEBUG REQUIRES: libomptarget-debug
#include <cassert>
#include <cstdio>
diff --git a/offload/test/offloading/ompx_coords.c b/offload/test/offloading/ompx_coords.c
index 61dad61f46..130b6f984f 100644
--- a/offload/test/offloading/ompx_coords.c
+++ b/offload/test/offloading/ompx_coords.c
@@ -26,7 +26,7 @@ int main(int argc, char **argv) {
int TL = 256;
int NT = (N + TL - 1) / TL;
-#pragma omp target data map(tofrom : X [0:N])
+#pragma omp target data map(tofrom : X[0 : N])
#pragma omp target teams num_teams(NT) thread_limit(TL)
{
#pragma omp parallel
diff --git a/offload/test/offloading/ompx_saxpy_mixed.c b/offload/test/offloading/ompx_saxpy_mixed.c
index 440b694e3a..defa9e3e1f 100644
--- a/offload/test/offloading/ompx_saxpy_mixed.c
+++ b/offload/test/offloading/ompx_saxpy_mixed.c
@@ -28,7 +28,7 @@ int main(int argc, char **argv) {
int TL = 256;
int NT = (N + TL - 1) / TL;
-#pragma omp target data map(to : X [0:N]) map(Y [0:N])
+#pragma omp target data map(to : X[0 : N]) map(Y[0 : N])
#pragma omp target teams num_teams(NT) thread_limit(TL)
{
#pragma omp parallel
diff --git a/offload/test/offloading/spmdization.c b/offload/test/offloading/spmdization.c
index 2cf30ff593..94a93b1dc3 100644
--- a/offload/test/offloading/spmdization.c
+++ b/offload/test/offloading/spmdization.c
@@ -29,14 +29,16 @@ int main(void) {
nested = omp_get_nested();
tid = omp_get_thread_num();
maxt = omp_get_max_threads();
- #pragma omp parallel
- noop();
+#pragma omp parallel
+ noop();
}
- printf("NumThreads: %i, InParallel: %i, Level: %i, ActiveLevel: %i, Nested: %i, "
- "ThreadNum: %i, MaxThreads: %i\n",
- nthreads, ip, lvl, alvl, nested, tid, maxt);
+ printf(
+ "NumThreads: %i, InParallel: %i, Level: %i, ActiveLevel: %i, Nested: %i, "
+ "ThreadNum: %i, MaxThreads: %i\n",
+ nthreads, ip, lvl, alvl, nested, tid, maxt);
// GENERIC: Generic mode
// SPMD: Generic-SPMD mode
- // CHECK: NumThreads: 1, InParallel: 0, Level: 0, ActiveLevel: 0, Nested: 0, ThreadNum: 0, MaxThreads:
+ // CHECK: NumThreads: 1, InParallel: 0, Level: 0, ActiveLevel: 0, Nested: 0,
+ // ThreadNum: 0, MaxThreads:
return 0;
}
diff --git a/offload/test/offloading/static_linking.c b/offload/test/offloading/static_linking.c
index 7be95a10ff..864ec4f294 100644
--- a/offload/test/offloading/static_linking.c
+++ b/offload/test/offloading/static_linking.c
@@ -1,6 +1,7 @@
// RUN: %libomptarget-compile-generic -DLIBRARY -c -o %t.o
// RUN: ar rcs %t.a %t.o
-// RUN: %libomptarget-compile-generic %t.a && %libomptarget-run-generic 2>&1 | %fcheck-generic
+// RUN: %libomptarget-compile-generic %t.a && %libomptarget-run-generic 2>&1 |
+// %fcheck-generic
#ifdef LIBRARY
int x = 42;
diff --git a/offload/test/offloading/target-tile.c b/offload/test/offloading/target-tile.c
index 8460b43b6f..2e67ed8ede 100644
--- a/offload/test/offloading/target-tile.c
+++ b/offload/test/offloading/target-tile.c
@@ -14,10 +14,10 @@
int main() {
int order[I_NTILES][J_NTILES][I_NELEMS][J_NELEMS];
int i, j;
- #pragma omp target map(tofrom: i, j)
+#pragma omp target map(tofrom : i, j)
{
int next = 0;
- #pragma omp tile sizes(I_NELEMS, J_NELEMS)
+#pragma omp tile sizes(I_NELEMS, J_NELEMS)
for (i = 0; i < I_NTILES * I_NELEMS; ++i) {
for (j = 0; j < J_NTILES * J_NELEMS; ++j) {
int iTile = i / I_NELEMS;
@@ -35,8 +35,8 @@ int main() {
for (int jElem = 0; jElem < J_NELEMS; ++jElem) {
int actual = order[iTile][jTile][iElem][jElem];
if (expected != actual) {
- printf("error: order[%d][%d][%d][%d] = %d, expected %d\n",
- iTile, jTile, iElem, jElem, actual, expected);
+ printf("error: order[%d][%d][%d][%d] = %d, expected %d\n", iTile,
+ jTile, iElem, jElem, actual, expected);
return 1;
}
++expected;
diff --git a/offload/test/offloading/target_constexpr_mapping.cpp b/offload/test/offloading/target_constexpr_mapping.cpp
index 14cf92a7cc..3148b64fb0 100644
--- a/offload/test/offloading/target_constexpr_mapping.cpp
+++ b/offload/test/offloading/target_constexpr_mapping.cpp
@@ -18,7 +18,7 @@ constexpr static double anotherPi = 3.14;
int main() {
double a[2];
-#pragma omp target map(tofrom : a[:2])
+#pragma omp target map(tofrom : a[ : 2])
{
a[0] = A::pi;
a[1] = anotherPi;
diff --git a/offload/test/offloading/target_critical_region.cpp b/offload/test/offloading/target_critical_region.cpp
index 9a741bef6c..a85c969560 100644
--- a/offload/test/offloading/target_critical_region.cpp
+++ b/offload/test/offloading/target_critical_region.cpp
@@ -22,9 +22,7 @@ int main() {
sum[0] = 0;
#pragma omp target teams distribute parallel for num_teams(256) \
- schedule(static, 1) map(to \
- : A[:N]) map(tofrom \
- : sum[:1])
+ schedule(static, 1) map(to : A[ : N]) map(tofrom : sum[ : 1])
{
for (int i = 0; i < N; i++) {
#pragma omp critical
diff --git a/offload/test/ompt/veccopy.c b/offload/test/ompt/veccopy.c
index 79cd918a60..6cc6ff782f 100644
--- a/offload/test/ompt/veccopy.c
+++ b/offload/test/ompt/veccopy.c
@@ -57,26 +57,35 @@ int main() {
/// CHECK: Callback Init:
/// CHECK: Callback Load:
/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
-/// CHECK: Callback Fini:
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 CHECK: Callback Fini:
diff --git a/offload/test/ompt/veccopy_disallow_both.c b/offload/test/ompt/veccopy_disallow_both.c
index 06293e413b..60b220894f 100644
--- a/offload/test/ompt/veccopy_disallow_both.c
+++ b/offload/test/ompt/veccopy_disallow_both.c
@@ -70,35 +70,27 @@ int main() {
/// CHECK-NOT: dest=(nil)
/// CHECK: Callback DataOp EMI: endpoint=1 optype=2
/// CHECK: Callback DataOp EMI: endpoint=2 optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback Target EMI: kind=1 endpoint=2
-/// CHECK: Callback Target EMI: kind=1 endpoint=1
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=1
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=1
-/// CHECK-NOT: dest=(nil)
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=2
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=3 CHECK: Callback DataOp EMI: endpoint=2 optype=3 CHECK:
+/// Callback DataOp EMI: endpoint=1 optype=3 CHECK: Callback DataOp EMI:
+/// endpoint=2 optype=3 CHECK: Callback DataOp EMI: endpoint=1 optype=4 CHECK:
+/// Callback DataOp EMI: endpoint=2 optype=4 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=4 CHECK: Callback DataOp EMI: endpoint=2 optype=4 CHECK:
+/// Callback Target EMI: kind=1 endpoint=2 CHECK: Callback Target EMI: kind=1
+/// endpoint=1 CHECK: Callback DataOp EMI: endpoint=1 optype=1 CHECK: Callback
+/// DataOp EMI: endpoint=2 optype=1 CHECK-NOT: dest=(nil) CHECK: Callback DataOp
+/// EMI: endpoint=1 optype=2 CHECK: Callback DataOp EMI: endpoint=2 optype=2
/// CHECK: Callback DataOp EMI: endpoint=1 optype=1
/// CHECK: Callback DataOp EMI: endpoint=2 optype=1
/// CHECK-NOT: dest=(nil)
/// CHECK: Callback DataOp EMI: endpoint=1 optype=2
/// CHECK: Callback DataOp EMI: endpoint=2 optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=3
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=1 optype=4
-/// CHECK: Callback DataOp EMI: endpoint=2 optype=4
-/// CHECK: Callback Target EMI: kind=1 endpoint=2
-/// CHECK: Callback Fini:
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=3 CHECK: Callback DataOp EMI: endpoint=2 optype=3 CHECK:
+/// Callback DataOp EMI: endpoint=1 optype=3 CHECK: Callback DataOp EMI:
+/// endpoint=2 optype=3 CHECK: Callback DataOp EMI: endpoint=1 optype=4 CHECK:
+/// Callback DataOp EMI: endpoint=2 optype=4 CHECK: Callback DataOp EMI:
+/// endpoint=1 optype=4 CHECK: Callback DataOp EMI: endpoint=2 optype=4 CHECK:
+/// Callback Target EMI: kind=1 endpoint=2 CHECK: Callback Fini:
diff --git a/offload/test/ompt/veccopy_map.c b/offload/test/ompt/veccopy_map.c
index 83f63a6e60..85e0094efd 100644
--- a/offload/test/ompt/veccopy_map.c
+++ b/offload/test/ompt/veccopy_map.c
@@ -54,31 +54,39 @@ int main() {
return rc;
}
-
/// CHECK: 0: Could not register callback 'ompt_callback_target_map'
/// CHECK: Callback Init:
/// CHECK: Callback Load:
/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
-/// CHECK: Callback Fini:
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 CHECK: Callback Fini:
diff --git a/offload/test/ompt/veccopy_no_device_init.c b/offload/test/ompt/veccopy_no_device_init.c
index c1a03191c3..f8872e29b3 100644
--- a/offload/test/ompt/veccopy_no_device_init.c
+++ b/offload/test/ompt/veccopy_no_device_init.c
@@ -54,26 +54,35 @@ int main() {
/// CHECK-NOT: Callback Init:
/// CHECK-NOT: Callback Load:
/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
-/// CHECK-NOT: Callback Fini:
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 CHECK-NOT: Callback Fini:
diff --git a/offload/test/ompt/veccopy_wrong_return.c b/offload/test/ompt/veccopy_wrong_return.c
index 2d07b4e1bf..d88ba65e10 100644
--- a/offload/test/ompt/veccopy_wrong_return.c
+++ b/offload/test/ompt/veccopy_wrong_return.c
@@ -54,26 +54,35 @@ int main() {
/// CHECK-NOT: Callback Init:
/// CHECK-NOT: Callback Load:
/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
-/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
-/// CHECK-NOT: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2
-/// CHECK-NOT: Callback Fini
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
+/// CHECK-NOT: Callback Submit: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3 CHECK-NOT: Callback DataOp:
+/// target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
+/// CHECK-NOT: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]]
+/// host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4 CHECK-NOT: Callback Target:
+/// target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 CHECK-NOT: Callback Fini
``````````
</details>
https://github.com/llvm/llvm-project/pull/75125
More information about the llvm-commits
mailing list