Skip to content

[PGO][Offload] Update PGO GPU tests #132262

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 5 commits into from
May 14, 2025
Merged

[PGO][Offload] Update PGO GPU tests #132262

Show file tree
Hide file tree
Changes from 1 commit
Commits
Show all changes
5 commits
Select commit Hold shift + click to select a range
10cb0f6
[PGO][Offload] Make PGO GPU tests atomic
EthanLuisMcDonough Mar 20, 2025
0c48346
[PGO][Offload] Update GPU PGO tests
EthanLuisMcDonough Apr 30, 2025
db194f2
Merge branch 'main' into gpu_pgo_atomic_tests2
EthanLuisMcDonough May 14, 2025
c200700
Merge branch 'main' into gpu_pgo_atomic_tests2
EthanLuisMcDonough May 14, 2025
5c5979c
[PGO][Offload] Change GPU PGO test names
EthanLuisMcDonough May 14, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
26 changes: 12 additions & 14 deletions offload/test/offloading/gpupgo/pgo1.c
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,13 +1,15 @@
// RUN: %libomptarget-compile-generic -fcreate-profile \
// RUN: -Xarch_device -fprofile-generate
// RUN: -Xarch_device -fprofile-generate \
// RUN: -Xarch_device -fprofile-update=atomic
// RUN: env LLVM_PROFILE_FILE=%basename_t.llvm.profraw \
// RUN: %libomptarget-run-generic 2>&1
// RUN: llvm-profdata show --all-functions --counts \
// RUN: %target_triple.%basename_t.llvm.profraw | \
// RUN: %fcheck-generic --check-prefix="LLVM-PGO"

// RUN: %libomptarget-compile-generic -fcreate-profile \
// RUN: -Xarch_device -fprofile-instr-generate
// RUN: -Xarch_device -fprofile-instr-generate \
// RUN: -Xarch_device -fprofile-update=atomic
// RUN: env LLVM_PROFILE_FILE=%basename_t.clang.profraw \
// RUN: %libomptarget-run-generic 2>&1
// RUN: llvm-profdata show --all-functions --counts \
Expand All @@ -23,18 +25,20 @@ int test2(int a) { return a * 2; }
int main() {
int m = 2;
#pragma omp target
for (int i = 0; i < 10; i++) {
m = test1(m);
for (int j = 0; j < 2; j++) {
m = test2(m);
{
for (int i = 0; i < 10; i++) {
m = test1(m);
for (int j = 0; j < 2; j++) {
m = test2(m);
}
}
}
}

// LLVM-PGO-LABEL: __omp_offloading_{{[_0-9a-zA-Z]*}}_main_{{[_0-9a-zA-Z]*}}:
// LLVM-PGO: Hash: {{0[xX][0-9a-fA-F]+}}
// LLVM-PGO: Counters: 4
// LLVM-PGO: Block counts: [20, 10, 2, 1]
// LLVM-PGO: Block counts: [20, 10, {{.*}}, 1]

// LLVM-PGO-LABEL: test1:
// LLVM-PGO: Hash: {{0[xX][0-9a-fA-F]+}}
Expand All @@ -53,14 +57,10 @@ int main() {
// LLVM-PGO-SAME: 3
// LLVM-PGO-LABEL: Maximum function count:
// LLVM-PGO-SAME: 20
// LLVM-PGO-LABEL: Maximum internal block count:
// LLVM-PGO-SAME: 10

// CLANG-PGO-LABEL: __omp_offloading_{{[_0-9a-zA-Z]*}}_main_{{[_0-9a-zA-Z]*}}:
// CLANG-PGO: Hash: {{0[xX][0-9a-fA-F]+}}
// CLANG-PGO: Counters: 3
// CLANG-PGO: Function count: 0
// CLANG-PGO: Block counts: [11, 20]
// CLANG-PGO: Block counts: [10, 20]

// CLANG-PGO-LABEL: test1:
// CLANG-PGO: Hash: {{0[xX][0-9a-fA-F]+}}
Expand All @@ -78,7 +78,5 @@ int main() {
// CLANG-PGO-SAME: Front-end
// CLANG-PGO-LABEL: Functions shown:
// CLANG-PGO-SAME: 3
// CLANG-PGO-LABEL: Maximum function count:
// CLANG-PGO-SAME: 20
// CLANG-PGO-LABEL: Maximum internal block count:
// CLANG-PGO-SAME: 20
24 changes: 14 additions & 10 deletions offload/test/offloading/gpupgo/pgo2.c
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,4 +1,5 @@
// RUN: %libomptarget-compile-generic -fprofile-generate
// RUN: %libomptarget-compile-generic -fprofile-generate \
// RUN: -fprofile-update=atomic
// RUN: env LLVM_PROFILE_FILE=%basename_t.llvm.profraw \
// RUN: %libomptarget-run-generic 2>&1
// RUN: llvm-profdata show --all-functions --counts \
Expand All @@ -8,7 +9,8 @@
// RUN: %target_triple.%basename_t.llvm.profraw \
// RUN: | %fcheck-generic --check-prefix="LLVM-DEVICE"

// RUN: %libomptarget-compile-generic -fprofile-instr-generate
// RUN: %libomptarget-compile-generic -fprofile-instr-generate \
// RUN: -fprofile-update=atomic
// RUN: env LLVM_PROFILE_FILE=%basename_t.clang.profraw \
// RUN: %libomptarget-run-generic 2>&1
// RUN: llvm-profdata show --all-functions --counts \
Expand All @@ -18,7 +20,8 @@
// RUN: %target_triple.%basename_t.clang.profraw | \
// RUN: %fcheck-generic --check-prefix="CLANG-DEV"

// RUN: %libomptarget-compile-generic -Xarch_host -fprofile-generate
// RUN: %libomptarget-compile-generic -Xarch_host -fprofile-generate \
// RUN: -fprofile-update=atomic
// RUN: env LLVM_PROFILE_FILE=%basename_t.nogpu.profraw \
// RUN: %libomptarget-run-generic 2>&1
// RUN: llvm-profdata show --all-functions --counts \
Expand All @@ -27,7 +30,7 @@
// RUN: not test -e %target_triple.%basename_t.nogpu.profraw

// RUN: %libomptarget-compile-generic -Xarch_host -fprofile-generate \
// RUN: -Xarch_device -fprofile-instr-generate
// RUN: -Xarch_device -fprofile-instr-generate -fprofile-update=atomic
// RUN: env LLVM_PROFILE_FILE=%basename_t.hidf.profraw \
// RUN: %libomptarget-run-generic 2>&1
// RUN: llvm-profdata show --all-functions --counts \
Expand All @@ -38,7 +41,7 @@
// RUN: | %fcheck-generic --check-prefix="CLANG-DEV"

// RUN: %libomptarget-compile-generic -Xarch_device -fprofile-generate \
// RUN: -Xarch_host -fprofile-instr-generate
// RUN: -Xarch_host -fprofile-instr-generate -fprofile-update=atomic
// RUN: env LLVM_PROFILE_FILE=%basename_t.hfdi.profraw \
// RUN: %libomptarget-run-generic 2>&1
// RUN: llvm-profdata show --all-functions --counts \
Expand All @@ -59,8 +62,10 @@ int main() {

int device_var = 1;
#pragma omp target
for (int i = 0; i < 10; i++) {
device_var *= i;
{
for (int i = 0; i < 10; i++) {
device_var *= i;
}
}
}

Expand All @@ -78,7 +83,7 @@ int main() {
// LLVM-DEVICE-LABEL: __omp_offloading_{{[_0-9a-zA-Z]*}}_main_{{[_0-9a-zA-Z]*}}:
// LLVM-DEVICE: Hash: {{0[xX][0-9a-fA-F]+}}
// LLVM-DEVICE: Counters: 3
// LLVM-DEVICE: Block counts: [10, 2, 1]
// LLVM-DEVICE: Block counts: [10, {{.*}}, 1]
// LLVM-DEVICE: Instrumentation level: IR

// CLANG-HOST-LABEL: main:
Expand All @@ -97,6 +102,5 @@ int main() {
// CLANG-DEV-LABEL: __omp_offloading_{{[_0-9a-zA-Z]*}}_main_{{[_0-9a-zA-Z]*}}:
// CLANG-DEV: Hash: {{0[xX][0-9a-fA-F]+}}
// CLANG-DEV: Counters: 2
// CLANG-DEV: Function count: 0
// CLANG-DEV: Block counts: [11]
// CLANG-DEV: Block counts: [10]
// CLANG-DEV: Instrumentation level: Front-end