maxjhandsome
/
llvm-project
forked from OSchip/llvm-project
1
0
Fork 0
Code Issues Pull Requests Packages Releases Wiki Activity

[OPENMP][NVPTX]Improve omp_get_max_threads() function. 

Summary:
Function omp_get_max_threads() can always return 1 if current execution
mode is SPMD.

Reviewers: grokos, gtbercea, kkwli0

Subscribers: guansong, jdoerfert, caomhin, openmp-commits

Tags: #openmp

Differential Revision: https://reviews.llvm.org/D61379

llvm-svn: 359792
This commit is contained in:
Alexey Bataev 2019-05-02 14:52:52 +00:00
parent 284472be6d
commit 4ad6dbc5fd
2 changed files with 48 additions and 5 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

7
openmp/libomptarget/deviceRTLs/nvptx/src/libcall.cu
View File

@ -54,14 +54,11 @@ EXTERN int omp_get_num_threads(void) {
}
EXTERN int omp_get_max_threads(void) {
if (isRuntimeUninitialized()) {
ASSERT0(LT_FUSSY, isSPMDMode(),
"Expected SPMD mode only with uninitialized runtime.");
if (isSPMDMode())
// We're already in parallel region.
return 1; // default is 1 thread avail
}
omptarget_nvptx_TaskDescr *currTaskDescr =
getMyTopTaskDescriptor(isSPMDMode());
getMyTopTaskDescriptor(/*isSPMDExecutionMode=*/false);
int rc = 1; // default is 1 thread avail
if (!currTaskDescr->InParallelRegion()) {
// Not currently in a parallel region, return what was set.

46
openmp/libomptarget/deviceRTLs/nvptx/test/api/max_threads.c Normal file
View File

@ -0,0 +1,46 @@
// RUN: %compile-run-and-check
#include <omp.h>
#include <stdio.h>
int main(int argc, char *argv[]) {
int MaxThreadsL1 = -1, MaxThreadsL2 = -1;
#pragma omp declare reduction(unique:int \
: omp_out = (omp_in == 1 ? omp_in : omp_out)) \
initializer(omp_priv = -1)
// Non-SPMD mode.
#pragma omp target teams map(MaxThreadsL1, MaxThreadsL2) thread_limit(32) \
num_teams(1)
{
MaxThreadsL1 = omp_get_max_threads();
#pragma omp parallel reduction(unique : MaxThreadsL2)
{ MaxThreadsL2 = omp_get_max_threads(); }
}
// CHECK: Non-SPMD MaxThreadsL1 = 32
printf("Non-SPMD MaxThreadsL1 = %d\n", MaxThreadsL1);
// CHECK: Non-SPMD MaxThreadsL2 = 1
printf("Non-SPMD MaxThreadsL2 = %d\n", MaxThreadsL2);
// SPMD mode with full runtime
MaxThreadsL2 = -1;
#pragma omp target parallel reduction(unique : MaxThreadsL2)
{ MaxThreadsL2 = omp_get_max_threads(); }
// CHECK: SPMD with full runtime MaxThreadsL2 = 1
printf("SPMD with full runtime MaxThreadsL2 = %d\n", MaxThreadsL2);
// SPMD mode without runtime
MaxThreadsL2 = -1;
#pragma omp target parallel for reduction(unique : MaxThreadsL2)
for (int I = 0; I < 2; ++I) {
MaxThreadsL2 = omp_get_max_threads();
}
// CHECK: SPMD without runtime MaxThreadsL2 = 1
printf("SPMD without runtime MaxThreadsL2 = %d\n", MaxThreadsL2);
return 0;
}