<div dir="ltr"><div class="gmail_quote"><div class="gmail_attr">Hi Ravi,<br></div><div dir="ltr" class="gmail_attr"><br></div><div dir="ltr" class="gmail_attr">On Tue, Aug 4, 2020 at 2:49 PM Narayanaswamy, Ravi <<a href="mailto:ravi.narayanaswamy@intel.com">ravi.narayanaswamy@intel.com</a>> wrote:<br></div><blockquote class="gmail_quote" style="margin:0px 0px 0px 0.8ex;border-left:1px solid rgb(204,204,204);padding-left:1ex">
<div lang="EN-US">
<div class="gmail-m_6872450235285706123WordSection1">
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">says it was intended that the reordering in L10-11 be applied first, and then the reordering in L14-16 be applied.
<u></u><u></u></span></p>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif"><u></u> <u></u></span></p>
<p class="MsoNormal">We probably need to discuss this in OpenMP Accel meeting, if you apply the above rule then present becomes a no-op.<u></u><u></u></p></div></div></blockquote><div><br></div><div>No, the present reordering rule is applied last, guaranteeing that map clauses with the present modifier are sorted first.<br></div><div><br></div><div>Joel<br></div><div><br></div><div> </div><blockquote class="gmail_quote" style="margin:0px 0px 0px 0.8ex;border-left:1px solid rgb(204,204,204);padding-left:1ex"><div lang="EN-US"><div class="gmail-m_6872450235285706123WordSection1"><p class="MsoNormal"></p>
<p class="MsoNormal"><u></u> <u></u></p>
<p class="MsoNormal">Ravi<u></u><u></u></p>
<p class="MsoNormal"><u></u> <u></u></p>
<p class="MsoNormal"><b>From:</b> Openmp-dev <<a href="mailto:openmp-dev-bounces@lists.llvm.org" target="_blank">openmp-dev-bounces@lists.llvm.org</a>> <b>
On Behalf Of </b>Joel E. Denny via Openmp-dev<br>
<b>Sent:</b> Tuesday, August 04, 2020 11:37 AM<br>
<b>To:</b> Openmp-dev <<a href="mailto:openmp-dev@lists.llvm.org" target="_blank">openmp-dev@lists.llvm.org</a>><br>
<b>Cc:</b> James Beyer <<a href="mailto:jbeyer@nvidia.com" target="_blank">jbeyer@nvidia.com</a>>; Tobias Burnus <<a href="mailto:tobias_burnus@mentor.com" target="_blank">tobias_burnus@mentor.com</a>><br>
<b>Subject:</b> [Openmp-dev] OpenMP TR8 present map type modifier issues<u></u><u></u></p>
<p class="MsoNormal"><u></u> <u></u></p>
<div>
<div>
<div>
<div>
<p class="MsoNormal">Hi,<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal">In the 7/29 OpenMP in LLVM call, we discussed several issues related to OpenMP TR8's present map type modifier. As suggested, I've discussed these issues with James Beyer. Below are the conclusions I gathered from that discussion and
my plans going forward. Please let me know if there are any concerns.<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<div>
<p class="MsoNormal"># Complaint for data that's not present on exit from target data region #<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal">OpenMP TR8 sec. 2.22.7.1 "map Clause", p. 321, L23-26 states:<u></u><u></u></p>
</div>
<div>
<div>
<div>
<blockquote style="border-color:currentcolor currentcolor currentcolor rgb(204,204,204);border-style:none none none solid;border-width:medium medium medium 1pt;padding:0in 0in 0in 6pt;margin-left:4.8pt;margin-right:0in">
<p>If the map clause appears on a target, target data, target enter<br>
data or target exit data construct with a present map-type-modifier<br>
then on entry to the region if the corresponding list item does not <br>
appear in the device data environment an error occurs and the <br>
program terminates.<u></u><u></u></p>
</blockquote>
<p class="MsoNormal">There is no corresponding statement about the exit from a region. As a result, the following test case is a valid OpenMP program that should not fail:<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<div>
<p class="MsoNormal" style="margin-left:30pt">
<span style="font-size:10pt;font-family:"Courier New"">#pragma omp target enter data map(alloc:i)<br>
#pragma omp target data map(present, alloc: i)<br>
{<br>
#pragma omp target exit data map(delete:i)</span><u></u><u></u></p>
</div>
<div style="margin-left:30pt">
<p class="MsoNormal"><span style="font-size:10pt;font-family:"Courier New"">} // i isn't present, but the present check doesn't happen on exit</span><u></u><u></u></p>
</div>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal">James agrees that this is the intended interpretation of TR8.<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal">My plan is to immediately push D84422, which implements this behavior and includes the above test.<u></u><u></u></p>
</div>
<div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<div>
<p class="MsoNormal"># Complaint for data that's not present on exit from target region #<u></u><u></u></p>
</div>
</div>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal">James agrees the above interpretation also applies to target regions. However, I have not for<span style="font-size:10pt">mulated a valid test case</span>. For example, James agrees the following attempt isn't valid OpenMP because it
has a race and might pass or fail:<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div style="margin-left:30pt">
<p class="MsoNormal"><span style="font-size:10pt;font-family:"Courier New"">#pragma omp target enter data map(alloc:i)</span><u></u><u></u></p>
</div>
<p class="MsoNormal" style="margin-left:30pt">
<span style="font-size:10pt;font-family:"Courier New"">#pragma omp target nowait map(present, alloc:i)</span><u></u><u></u></p>
<p class="MsoNormal" style="margin-left:30pt">
<span style="font-size:10pt;font-family:"Courier New"">// i might not be present here, so the entry present check might fail</span><u></u><u></u></p>
<p class="MsoNormal" style="margin-left:30pt">
<span style="font-size:10pt;font-family:"Courier New"">{}</span><u></u><u></u></p>
<p class="MsoNormal" style="margin-left:30pt">
<span style="font-size:10pt;font-family:"Courier New"">#pragma omp target exit data map(delete:i)</span><u></u><u></u></p>
<p class="MsoNormal"><u></u> <u></u></p>
<p class="MsoNormal"><span style="font-size:10pt;font-family:"Arial",sans-serif">D84422 includes a comment saying we're not bothering to eliminate the presence check on exit from a target region because
we lack a test case.</span><u></u><u></u></p>
<p class="MsoNormal"><u></u> <u></u></p>
<p class="MsoNormal"><span style="font-size:10pt;font-family:"Arial",sans-serif">My plan is not to pursue this issue for now. If someone later manages to write a test, I'd be happy to fix the implementation
in a followup patch.</span><u></u><u></u></p>
<p class="MsoNormal"><u></u> <u></u></p>
<div>
<p class="MsoNormal"># Effective ordering of map clauses #<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal">OpenMP TR8 sec. 2.22.7.1 "map Clause", p. 319, L10-11 states:<u></u><u></u></p>
</div>
<div>
<div>
<blockquote style="border-color:currentcolor currentcolor currentcolor rgb(204,204,204);border-style:none none none solid;border-width:medium medium medium 1pt;padding:0in 0in 0in 6pt;margin-left:4.8pt;margin-right:0in">
<p>For a given construct, the effect of a map clause with the to, from,<br>
or tofrom map-type is ordered before the effect of a map clause with<br>
the alloc, release, or delete map-type.<u></u><u></u></p>
</blockquote>
</div>
</div>
<div>
<p class="MsoNormal">However, L14-16 states:<u></u><u></u></p>
</div>
<div>
<blockquote style="border-color:currentcolor currentcolor currentcolor rgb(204,204,204);border-style:none none none solid;border-width:medium medium medium 1pt;padding:0in 0in 0in 6pt;margin-left:4.8pt;margin-right:0in">
<p>If a map clause with a present map-type-modifier is present in a map <br>
clause, then the effect of the clause is ordered before all other<br>
map clauses that do not have the present modifier.<u></u><u></u></p>
</blockquote>
<p class="MsoNormal">James agrees this combination is contradictory for cases like<span style="font-family:"Arial",sans-serif"> `map(tofrom:var) map(present,alloc:var)`. He says it was intended that the reordering in L10-11 be applied first, and then the reordering
in L14-16 be applied. We also discussed with Tobias Burnus from Mentor, who then filed OpenMP spec issue 2337, which also covers additional issues with the way reordering is specified.</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">My plan for this issue is to add the following entries to Clang's OpenMP status doc:</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">For 5.0, in which L10-11 also appear but do not seem to be implemented in Clang:</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">```</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">| device extension | map clause reordering based on map types | :none:`unclaimed` |
|</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">```</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">For 5.1:</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><u></u> <u></u></p>
</div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">```</span><u></u><u></u></p>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">| device extension | map clause reordering reordering based on 'present' modifier | :none:`unclaimed` |
|</span><u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="font-family:"Arial",sans-serif">```</span><u></u><u></u></p>
</div>
</div>
</div>
<p class="MsoNormal"><u></u> <u></u></p>
<div>
<p class="MsoNormal">Thanks.<u></u><u></u></p>
</div>
<div>
<p class="MsoNormal"><span style="color:rgb(136,136,136)"><u></u> <u></u></span></p>
</div>
<div>
<p class="MsoNormal"><span style="color:rgb(136,136,136)">Joel<u></u><u></u></span></p>
</div>
</div>
</div>
</div>
<div>
<p class="MsoNormal" style="margin-bottom:12pt"><u></u> <u></u></p>
</div>
</div>
</div>
</div>
</blockquote></div></div>