Page MenuHomePhabricator

gtbercea (Gheorghe-Teodor Bercea)
User

Projects

User does not belong to any projects.

User Details

User Since
Dec 29 2016, 12:44 AM (129 w, 5 d)

Recent Activity

Thu, Jun 20

gtbercea committed rG0034e84aa542: [OpenMP] Add support for handling declare target to clause when unified memory… (authored by gtbercea).
[OpenMP] Add support for handling declare target to clause when unified memory…
Thu, Jun 20, 11:02 AM
gtbercea updated the diff for D63108: [OpenMP] Add support for handling declare target to clause when unified memory is required.
  • Address comments.
Thu, Jun 20, 9:26 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D63108: [OpenMP] Add support for handling declare target to clause when unified memory is required.
  • Address comments.
Thu, Jun 20, 8:45 AM · Restricted Project, Restricted Project

Wed, Jun 19

gtbercea added a comment to D63108: [OpenMP] Add support for handling declare target to clause when unified memory is required.

Still need to update the test but the rest of the code is updated.

Wed, Jun 19, 2:12 PM · Restricted Project, Restricted Project
gtbercea updated the diff for D63108: [OpenMP] Add support for handling declare target to clause when unified memory is required.
  • Merge MT_Link and MT_To with unified memory cases.
  • Transform switch into if statements.
  • Fix declare target attribute checks.
Wed, Jun 19, 2:09 PM · Restricted Project, Restricted Project
gtbercea closed D63009: [OpenMP] Add target task alloc function with device ID.
Wed, Jun 19, 12:32 PM · Restricted Project, Restricted Project, Restricted Project
gtbercea committed rGaace6d285d7c: [OpenMP][libomptarget] Add support for declare target to clause under unified… (authored by gtbercea).
[OpenMP][libomptarget] Add support for declare target to clause under unified…
Wed, Jun 19, 8:47 AM
gtbercea added inline comments to D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.
Wed, Jun 19, 8:42 AM · Restricted Project, Restricted Project
gtbercea added inline comments to D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.
Wed, Jun 19, 8:02 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.
  • Fix condition.
Wed, Jun 19, 7:59 AM · Restricted Project, Restricted Project
gtbercea committed rGc74707cb3904: [OpenMP] Strengthen regression tests for task allocation under nowait depend… (authored by gtbercea).
[OpenMP] Strengthen regression tests for task allocation under nowait depend…
Wed, Jun 19, 7:27 AM
gtbercea added a comment to D63009: [OpenMP] Add target task alloc function with device ID.

Fixed in D63454

Wed, Jun 19, 7:24 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea added a comment to D63454: [OpenMP] Strengthen regression tests for task allocation under nowait depend clauses NFC.

ping

Wed, Jun 19, 6:38 AM · Restricted Project, Restricted Project

Mon, Jun 17

gtbercea updated the diff for D63454: [OpenMP] Strengthen regression tests for task allocation under nowait depend clauses NFC.
  • Fix test.
Mon, Jun 17, 2:19 PM · Restricted Project, Restricted Project
gtbercea added a reviewer for D63454: [OpenMP] Strengthen regression tests for task allocation under nowait depend clauses NFC: jdoerfert.
Mon, Jun 17, 12:53 PM · Restricted Project, Restricted Project
gtbercea created D63454: [OpenMP] Strengthen regression tests for task allocation under nowait depend clauses NFC.
Mon, Jun 17, 12:52 PM · Restricted Project, Restricted Project

Sun, Jun 16

gtbercea added a comment to D63009: [OpenMP] Add target task alloc function with device ID.

Am I correct that the second to last revision ("- Fix tests.") removed all checks for the actual device_id argument from the tests? From my point of view that's not fixing but weakening the tests! Can you explain why they needed "fixing"?

When I was just passing the default value the LLVM-IR was: i64 -1 i.e. constant, easy to check.

With the latest change the emitted code is: i64 %123 i.e. where %123 is a local derived from the expression of the device ID.

If the value is constant, check for the constant.

Sun, Jun 16, 10:08 AM · Restricted Project, Restricted Project, Restricted Project

Sat, Jun 15

gtbercea added a comment to D63009: [OpenMP] Add target task alloc function with device ID.

Am I correct that the second to last revision ("- Fix tests.") removed all checks for the actual device_id argument from the tests? From my point of view that's not fixing but weakening the tests! Can you explain why they needed "fixing"?

Sat, Jun 15, 8:28 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea added a comment to D63009: [OpenMP] Add target task alloc function with device ID.

The tests must check the device ID for target-based calls of the task alloc function.

Sat, Jun 15, 8:22 AM · Restricted Project, Restricted Project, Restricted Project

Fri, Jun 14

gtbercea committed rG545a9fe1063d: [OpenMP] Add target task alloc function with device ID (authored by gtbercea).
[OpenMP] Add target task alloc function with device ID
Fri, Jun 14, 1:19 PM
gtbercea added inline comments to D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.
Fri, Jun 14, 1:19 PM · Restricted Project, Restricted Project
gtbercea committed rGb48e44a65cff: [OpenMP] Add task alloc function (authored by gtbercea).
[OpenMP] Add task alloc function
Fri, Jun 14, 1:13 PM
gtbercea added inline comments to D63009: [OpenMP] Add target task alloc function with device ID.
Fri, Jun 14, 12:14 PM · Restricted Project, Restricted Project, Restricted Project
gtbercea added inline comments to D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.
Fri, Jun 14, 11:24 AM · Restricted Project, Restricted Project
gtbercea added a comment to D63010: [OpenMP] Add task alloc function.

ping

Fri, Jun 14, 11:13 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea committed rG5254f0a9abab: [OpenMP] Avoid emitting maps for target link variables when unified memory is… (authored by gtbercea).
[OpenMP] Avoid emitting maps for target link variables when unified memory is…
Fri, Jun 14, 10:56 AM
gtbercea updated the diff for D60883: [OpenMP] Avoid emitting maps for target link variables when unified memory is used.
  • Make function const.
Fri, Jun 14, 10:37 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D60883: [OpenMP] Avoid emitting maps for target link variables when unified memory is used.
  • Remove virtual.
Fri, Jun 14, 9:02 AM · Restricted Project, Restricted Project
gtbercea added a comment to D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.

ping

Fri, Jun 14, 8:58 AM · Restricted Project, Restricted Project
gtbercea added a comment to D63108: [OpenMP] Add support for handling declare target to clause when unified memory is required.

ping

Fri, Jun 14, 8:58 AM · Restricted Project, Restricted Project
gtbercea added inline comments to D63009: [OpenMP] Add target task alloc function with device ID.
Fri, Jun 14, 8:52 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea added inline comments to D63009: [OpenMP] Add target task alloc function with device ID.
Fri, Jun 14, 7:50 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea added inline comments to D63009: [OpenMP] Add target task alloc function with device ID.
Fri, Jun 14, 7:37 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea updated the diff for D63009: [OpenMP] Add target task alloc function with device ID.
  • Fix function name.
Fri, Jun 14, 7:14 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea added inline comments to D63009: [OpenMP] Add target task alloc function with device ID.
Fri, Jun 14, 7:14 AM · Restricted Project, Restricted Project, Restricted Project

Wed, Jun 12

gtbercea abandoned D38978: [OpenMP] Enable the lowering of implicitly shared variables in OpenMP GPU-offloaded target regions to the GPU shared memory.

Alternative solution was implemented.

Wed, Jun 12, 10:16 AM
gtbercea abandoned D51446: [OpenMP][bugfix] Add missing macros for Power.
Wed, Jun 12, 10:13 AM · Restricted Project
gtbercea updated the diff for D63009: [OpenMP] Add target task alloc function with device ID.
  • Fix tests.
Wed, Jun 12, 10:09 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea updated the diff for D63009: [OpenMP] Add target task alloc function with device ID.
  • Add device ID if available.
Wed, Jun 12, 8:01 AM · Restricted Project, Restricted Project, Restricted Project

Tue, Jun 11

gtbercea updated the diff for D63009: [OpenMP] Add target task alloc function with device ID.
  • Add tests.
Tue, Jun 11, 1:10 PM · Restricted Project, Restricted Project, Restricted Project
gtbercea updated the diff for D63010: [OpenMP] Add task alloc function.
  • Add temporary implementation.
Tue, Jun 11, 1:08 PM · Restricted Project, Restricted Project, Restricted Project
gtbercea updated the diff for D63009: [OpenMP] Add target task alloc function with device ID.
  • Add temporary implementation.
Tue, Jun 11, 1:05 PM · Restricted Project, Restricted Project, Restricted Project
gtbercea updated the diff for D63010: [OpenMP] Add task alloc function.
  • Add tests.
Tue, Jun 11, 1:02 PM · Restricted Project, Restricted Project, Restricted Project

Mon, Jun 10

gtbercea updated the summary of D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.
Mon, Jun 10, 4:48 PM · Restricted Project, Restricted Project
gtbercea created D63108: [OpenMP] Add support for handling declare target to clause when unified memory is required.
Mon, Jun 10, 4:48 PM · Restricted Project, Restricted Project
gtbercea created D63106: [OpenMP][libomptarget] Add support for declare target to clause under unified memory.
Mon, Jun 10, 4:43 PM · Restricted Project, Restricted Project

Fri, Jun 7

gtbercea created D63010: [OpenMP] Add task alloc function.
Fri, Jun 7, 9:05 AM · Restricted Project, Restricted Project, Restricted Project
gtbercea created D63009: [OpenMP] Add target task alloc function with device ID.
Fri, Jun 7, 8:57 AM · Restricted Project, Restricted Project, Restricted Project

Tue, Jun 4

gtbercea committed rGc5fe030c166b: [OpenMP][libomptarget] Enable usage of unified memory for declare target link… (authored by gtbercea).
[OpenMP][libomptarget] Enable usage of unified memory for declare target link…
Tue, Jun 4, 8:05 AM

Mon, Jun 3

gtbercea added a comment to D59319: [OpenMP][Offloading][1/3] A generic and simple target region interface.

Could you check what the difference is between the same kernel in today's SPMD mode vs the SPMD mode produced via this method? Number of registers, instructions, checking everything gets optimized out as expected. The LLVM-IR should be almost identical.

Mon, Jun 3, 12:55 PM · Restricted Project
gtbercea added a comment to D59319: [OpenMP][Offloading][1/3] A generic and simple target region interface.

Could you add some tests for this?

Mon, Jun 3, 12:37 PM · Restricted Project
gtbercea added a comment to D59424: [OpenMP][NVPTX] Replace void** buffer by byte-wise buffer.

Could you check if there is any change in the number of registers new scheme vs. old scheme?

Mon, Jun 3, 12:28 PM · Restricted Project
gtbercea added inline comments to D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
Mon, Jun 3, 12:04 PM · Restricted Project, Restricted Project
gtbercea added inline comments to D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
Mon, Jun 3, 11:37 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
  • Clean-up code.
Mon, Jun 3, 7:39 AM · Restricted Project, Restricted Project
gtbercea added inline comments to D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
Mon, Jun 3, 7:19 AM · Restricted Project, Restricted Project

Wed, May 29

gtbercea added a comment to D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.

ping

Wed, May 29, 8:44 AM · Restricted Project, Restricted Project

Tue, May 28

gtbercea added a comment to D56274: [OPENMP][NVPTX]Fix incompatibility of __syncthreads with LLVM, NFC..

The optimization that is being applied is called "call site splitting" in LLVM.

Tue, May 28, 12:33 PM · Restricted Project
gtbercea added a comment to D56274: [OPENMP][NVPTX]Fix incompatibility of __syncthreads with LLVM, NFC..

Is OpenMP not marking all functions as convergent?

ping

Marks,but some of the optimizations ignore this attribute. I don't remebet which one exactly, something like critical edge splitting.

I think critical edge splitting handles convergent correctly, since it is one of the motivating examples. I just looked at a random example in test/OpenMP, and this doesn't look correct to me:

__kmpc_barrier is declared as convergent, but the callers are not:

declare void @__kmpc_barrier(%struct.ident_t*, i32) #3
define internal void @__omp_outlined__78(i32* noalias %.global_tid., i32* noalias %.bound_tid.) #0 {
attributes #0 = { noinline norecurse nounwind optnone "correctly-rounded-divide-sqrt-fp-math"="false" "disable-tail-calls"="false" "less-precise-fpmad"="false" "min-legal-vector-width"="0" "no-frame-pointer-elim"="false" "no-infs-fp-math"="false" "no-jump-tables"="false" "no-nans-fp-math"="false" "no-signed-zeros-fp-math"="false" "no-trapping-math"="false" "stack-protector-buffer-size"="8" "target-features"="+ptx32,+sm_20" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #1 = { nounwind readnone }
attributes #2 = { argmemonly nounwind }

*All* functions need to be assumed convergent, not just the convergent barrier leafs.

The problem is not in the OpenMP code, it is in Cuda code. It appears only when we inline the runtime written in Cuda, where everything is marked correctly. For OpenMP code it is not necessary to mark all the functions as convergent, all required functions are marked by Cuda.

I don't follow how this is unnecessary. This is producing an IR module with a convergent call from a non-convergent function. This is plainly broken, and the verifier should probably reject it. Any transform on the caller of these could violate the convergent rules. The IR should be semantically correct at all times regardless of what is inlined or linked

+1 to the verifier check. @jlebar , do you agree?

If the verifier is broken, it must be fixed, of course. and kmpc_barrier too. But the problem still remains. One of the functions, at least, that calculates cost of the function in splitting edge, does not take convergent attribute into account and it leads to dangerous optimizations.

Is there a public test case? If not, can you share/construct one?

Better to ask Doru, he tried to investigate this problem (after my patch, which is just a copy of the named barriers, asm volatile construct does not have this problem) and, if I recall it correctly, reported about this problem. But I'm not sure, to whom he reported, to LLVM or to NVidia.

I reported several problems to NVIDIA. Is the problem below the one you're referring to?

For the following CUDA code:

if (threadIdx.x == 0) {
// do some initialization (A)
}
__synchtreads();
// some code (B)

when I enable optimizations I get the syncthreads being duplicated and the code hangs at runtime:

entry:
%0 = tail call i32 @llvm.nvvm.read.ptx.sreg.tid.x() #6, !range !12
%cmp.i2 = icmp eq i32 %0, 0
br i1 %cmp.i2, label %if.then, label %if.end.split

if.end.split:
tail call void @llvm.nvvm.barrier0()#6
// LLVM IR for B code block
br label %if.end

if.then:
// LLVM IR for A code block
tail call void @llvm.nvvm.barrier0()#6
// LLVM IR for B code block
br label %if.end

if.end:

Can you post the starting IR for this?

Tue, May 28, 12:27 PM · Restricted Project
gtbercea added a comment to D56274: [OPENMP][NVPTX]Fix incompatibility of __syncthreads with LLVM, NFC..

Is OpenMP not marking all functions as convergent?

ping

Marks,but some of the optimizations ignore this attribute. I don't remebet which one exactly, something like critical edge splitting.

I think critical edge splitting handles convergent correctly, since it is one of the motivating examples. I just looked at a random example in test/OpenMP, and this doesn't look correct to me:

__kmpc_barrier is declared as convergent, but the callers are not:

declare void @__kmpc_barrier(%struct.ident_t*, i32) #3
define internal void @__omp_outlined__78(i32* noalias %.global_tid., i32* noalias %.bound_tid.) #0 {
attributes #0 = { noinline norecurse nounwind optnone "correctly-rounded-divide-sqrt-fp-math"="false" "disable-tail-calls"="false" "less-precise-fpmad"="false" "min-legal-vector-width"="0" "no-frame-pointer-elim"="false" "no-infs-fp-math"="false" "no-jump-tables"="false" "no-nans-fp-math"="false" "no-signed-zeros-fp-math"="false" "no-trapping-math"="false" "stack-protector-buffer-size"="8" "target-features"="+ptx32,+sm_20" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #1 = { nounwind readnone }
attributes #2 = { argmemonly nounwind }

*All* functions need to be assumed convergent, not just the convergent barrier leafs.

The problem is not in the OpenMP code, it is in Cuda code. It appears only when we inline the runtime written in Cuda, where everything is marked correctly. For OpenMP code it is not necessary to mark all the functions as convergent, all required functions are marked by Cuda.

I don't follow how this is unnecessary. This is producing an IR module with a convergent call from a non-convergent function. This is plainly broken, and the verifier should probably reject it. Any transform on the caller of these could violate the convergent rules. The IR should be semantically correct at all times regardless of what is inlined or linked

+1 to the verifier check. @jlebar , do you agree?

If the verifier is broken, it must be fixed, of course. and kmpc_barrier too. But the problem still remains. One of the functions, at least, that calculates cost of the function in splitting edge, does not take convergent attribute into account and it leads to dangerous optimizations.

Is there a public test case? If not, can you share/construct one?

Better to ask Doru, he tried to investigate this problem (after my patch, which is just a copy of the named barriers, asm volatile construct does not have this problem) and, if I recall it correctly, reported about this problem. But I'm not sure, to whom he reported, to LLVM or to NVidia.

Tue, May 28, 12:11 PM · Restricted Project
gtbercea added inline comments to D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
Tue, May 28, 11:24 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
  • Change return type to 64 bits.
Tue, May 28, 11:23 AM · Restricted Project, Restricted Project

May 24 2019

gtbercea added a comment to D60883: [OpenMP] Avoid emitting maps for target link variables when unified memory is used.

Just one question after looking at the test: how we're going to link the device variable to its host copy? I think the address should be passed to the runtime, no?

May 24 2019, 5:34 PM · Restricted Project, Restricted Project
gtbercea added a comment to D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.

ping

May 24 2019, 12:12 PM · Restricted Project, Restricted Project
gtbercea updated the diff for D60883: [OpenMP] Avoid emitting maps for target link variables when unified memory is used.
  • Add test.
May 24 2019, 12:09 PM · Restricted Project, Restricted Project
gtbercea committed rG93d2c8a646c5: [OpenMP] Add test for requires and unified shared memory clause with declare… (authored by gtbercea).
[OpenMP] Add test for requires and unified shared memory clause with declare…
May 24 2019, 11:49 AM
gtbercea added a comment to D62407: [OpenMP] Add test for requires and unified shared memory clause with declare target link.

Just one question - why it is not the part of the whole patch for the unified memory support in clang?

May 24 2019, 10:53 AM · Restricted Project
gtbercea created D62407: [OpenMP] Add test for requires and unified shared memory clause with declare target link.
May 24 2019, 9:55 AM · Restricted Project

May 22 2019

gtbercea updated the diff for D60883: [OpenMP] Avoid emitting maps for target link variables when unified memory is used.
  • Include sema check.
May 22 2019, 2:10 PM · Restricted Project, Restricted Project
gtbercea added inline comments to D60883: [OpenMP] Avoid emitting maps for target link variables when unified memory is used.
May 22 2019, 10:06 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D60883: [OpenMP] Avoid emitting maps for target link variables when unified memory is used.
  • Fix function call.
May 22 2019, 10:05 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
  • Remove function from generic plugin.
May 22 2019, 8:59 AM · Restricted Project, Restricted Project

May 21 2019

gtbercea committed rG66cdbb47d2f8: [OpenMP] Add support for registering requires directives with the runtime (authored by gtbercea).
[OpenMP] Add support for registering requires directives with the runtime
May 21 2019, 12:41 PM
gtbercea committed rG9e9c918259f6: [OpenMP][libomptarget] Enable requires flags for target libraries. (authored by gtbercea).
[OpenMP][libomptarget] Enable requires flags for target libraries.
May 21 2019, 12:33 PM
gtbercea added a comment to D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.

Moved some of the functionality from D60223 to here since it fits better with what this patch is trying to achieve.

May 21 2019, 11:49 AM · Restricted Project, Restricted Project
gtbercea added a reviewer for D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables: grokos.
May 21 2019, 11:47 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D60884: [OpenMP][libomptarget] Enable usage of unified memory for declare target link variables.
  • Disable device copies for target link variables under unified memory.
  • Add requires init on plugin side.
May 21 2019, 11:47 AM · Restricted Project, Restricted Project
gtbercea added a comment to D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..

@Hahnfeld just updated the patch. Please let me know if you have any more comments.

May 21 2019, 11:33 AM · Restricted Project
gtbercea updated the diff for D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..
  • Remove requires init on plugin side.
May 21 2019, 11:31 AM · Restricted Project
gtbercea updated the diff for D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..
  • Format test.
May 21 2019, 8:55 AM · Restricted Project
gtbercea added a comment to D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..

In this patch __tgt_rtl_init_requires does nothing. Do we really want to add the prototype now? If yes, are we sure that we won't need to pass more information in the future? Once we add it, the runtime library must maintain compatibility.

Ping. At the moment this patch just stores the RequiresFlags in the plugins and never uses it.

May 21 2019, 8:51 AM · Restricted Project
gtbercea updated the diff for D60568: [OpenMP] Add support for registering requires directives with the runtime.
  • Remove new option.
May 21 2019, 7:26 AM · Restricted Project, Restricted Project
gtbercea added inline comments to D60568: [OpenMP] Add support for registering requires directives with the runtime.
May 21 2019, 7:01 AM · Restricted Project, Restricted Project

May 20 2019

gtbercea added a comment to D60568: [OpenMP] Add support for registering requires directives with the runtime.

Looks good, in general. Just one more question: will tgt_register_requires be emitted if only -fopenmp is used? If -fopenmp-targets is not provided, this function should not be called, I think, because we're not going to use offloading at all in this case and tgt_register_requires should not be emitted. Otherwise, we may end up with the linker error that tgt_register_requires is not found. Would be good to have a test with -fopenmp and #pragma omp requires that check that no tgt_register_requires is emitted in this case.

May 20 2019, 2:07 PM · Restricted Project, Restricted Project
gtbercea updated the diff for D60568: [OpenMP] Add support for registering requires directives with the runtime.
  • Avoid requires registration when no targets are specified.
May 20 2019, 2:06 PM · Restricted Project, Restricted Project
gtbercea added a comment to D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..

In this patch __tgt_rtl_init_requires does nothing. Do we really want to add the prototype now? If yes, are we sure that we won't need to pass more information in the future? Once we add it, the runtime library must maintain compatibility.

Moreover, I'm not sure if we should require plugins to implement the function if they don't need any addition checks. I'd propose to make this optional, ie look for the symbol via dlsym and disable the call if it's not found.

May 20 2019, 12:29 PM · Restricted Project
gtbercea updated the diff for D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..
  • Make implementation of init requires optional.
May 20 2019, 12:29 PM · Restricted Project
gtbercea added a comment to D60568: [OpenMP] Add support for registering requires directives with the runtime.

ping

May 20 2019, 10:27 AM · Restricted Project, Restricted Project
gtbercea updated the diff for D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..
  • Fix test to make it compatible with previous clang versions.
May 20 2019, 10:10 AM · Restricted Project

May 17 2019

gtbercea updated the diff for D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..
  • Add comment.
May 17 2019, 3:38 PM · Restricted Project
gtbercea updated the diff for D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..
  • Fix test.
May 17 2019, 3:27 PM · Restricted Project
gtbercea added inline comments to D60223: [OpenMP][libomptarget] Enable requires flags for target libraries..
May 17 2019, 1:55 PM · Restricted Project
gtbercea updated the diff for D60568: [OpenMP] Add support for registering requires directives with the runtime.
  • Fix braces.
May 17 2019, 1:37 PM · Restricted Project, Restricted Project
gtbercea updated the diff for D60568: [OpenMP] Add support for registering requires directives with the runtime.
  • Fix format.
May 17 2019, 1:34 PM · Restricted Project, Restricted Project
gtbercea updated the diff for D60568: [OpenMP] Add support for registering requires directives with the runtime.
  • Fix conditions.
May 17 2019, 1:31 PM · Restricted Project, Restricted Project
gtbercea committed rG144291e14c1b: [OpenMP][bugfix] Add missing math functions variants for log and abs. (authored by gtbercea).
[OpenMP][bugfix] Add missing math functions variants for log and abs.
May 17 2019, 12:14 PM
gtbercea added inline comments to D60568: [OpenMP] Add support for registering requires directives with the runtime.
May 17 2019, 8:22 AM · Restricted Project, Restricted Project
gtbercea added a comment to D62046: [OpenMP][bugfix] Add missing math functions variants for log and abs..

@tra I eliminated the long double definition for log and only left the declaration.

May 17 2019, 8:21 AM · Restricted Project
gtbercea updated the diff for D62046: [OpenMP][bugfix] Add missing math functions variants for log and abs..
  • update patch
May 17 2019, 8:19 AM · Restricted Project
gtbercea added inline comments to D60568: [OpenMP] Add support for registering requires directives with the runtime.
May 17 2019, 8:03 AM · Restricted Project, Restricted Project