Would someone be able to commit this? Thanks a lot.

• Simplify target region.

• Rebase.

• Add back requires registration call.

In D65340#1623325, @Hahnfeld wrote:

I disagree, the information about requires directives only needs to be present when initializing the device, not when loading a library. I've posted D66019 to solve this and fix the test I mentioned in D65001.

I see what you mean! I was working on a patch for adding a new method for manually setting the flag but this works too if we don't have to have a per device flag.

@Hahnfeld
I would like to clarify something about tgt_register_requires call.

• Remove requires manual call.

In D65340#1621626, @grokos wrote:

I don't have any comments about this patch other than the ones @Hahnfeld mentioned.

@Hahnfeld @grokos any further comments?

@grokos I have updated this patch as suggested and have added new tests. Please let me know if you any further comments.

• Update.

• Clean up conditions.

• Move condition.

@grokos any thoughts on the updated condition?

@Hahnfeld I have updated the tests, I added two new ones. Let me know if you have any further comments.

• Clean test.

• Clean-up.

• Add unsupported clang versions to test.
• Add more tests.

In D65340#1616793, @Hahnfeld wrote:

The test will currently fail with older versions of Clang. It must at least be marked UNSUPPORTED for Clang versions older than what-will-be Clang 10.

Additionally, I'd still like to see a small test with manual calls to __tgt_target_data_begin / _end if possible. I'm thinking of the following:

int *a = // malloc
int *device;

__tgt_target_data_begin(...); // with close modifier for a
#pragma omp target data use_device_ptr(a)
{
  device = a;
}
__tgt_target_data_end(...); // correspondingly to the begin call above

// check...

Instead of the nested target data region, you could also use a simple target region in the outer data region, and have a check for use_device_ptr in another test.

@Hahnfeld I have addressed your comments, please let me know if you have further comments.

• Remove close flag from checks where it is not needed.

• Add back close modifier flag check.

• Format test.

@grokos I have updated the condition. I have also amended the comment to better explain the conditions.

• Format tests.

• Fix condition to exclude already mapped device data.

• Fix condition again.

• Fix condition.

• Update.

• Clean-up.

• Address comments.

• Add manual for registering the requires flag.
• Address comments.

In D65001#1610445, @Hahnfeld wrote:

There's still no call to __tgt_register_requires in the two tests, so I guess they won't pass with older versions of Clang.

• Add close modifier flag.
• Move test to appropriate folder.

@Hahnfeld I have merged the tests from the previous patch as per Alexey's suggestion - with minor changes to make them pass. Let me know if this now addresses your previous comments.

• Improve test.
• Add tests from previous patch.

@Hahnfeld I hope I have answered all your concerns. Please let me know if you have any other comments.

• Rename test.
• Revert file move. Simplify condition.
• Address comments.

@Hahnfeld @grokos new version of the patch with addressed comments.

• Move tests to new folder.
• Fix tests.
• Move requires test.

• Address comments.

• Remove use device ptr flag.

In D65341#1604562, @kkwli0 wrote:

In D65341#1604561, @gtbercea wrote:

In D65341#1604440, @ABataev wrote:

In D65341#1604411, @Hahnfeld wrote:

There's already D55892 with a better set of tests, including target enter data / target exit data.

Better to merge those two patches into one.

How would you like me to proceed? It looks like the other patch has been sitting there approved for many many months.

Yes, the patch has been waiting for the runtime implementation first and then commit.

In D65341#1604440, @ABataev wrote:

In D65341#1604411, @Hahnfeld wrote:

There's already D55892 with a better set of tests, including target enter data / target exit data.

Better to merge those two patches into one.

In D65340#1604437, @Hahnfeld wrote:

I think this mostly looks good, but depends on D65001, right?

• Improve test.

In D65001#1601037, @Hahnfeld wrote:

In D65001#1600995, @gtbercea wrote:

@Hahnfeld I added several tests. Because these tests require unified memory to be supported by the underlying system I have added them as a new type of test: check-libomptarget-nvptx-unified (in addition to the check-libomptarget-nvptx one). These tests should only be run on platforms which support unified memory.

Can we instead have the tests work with the host plugins? That would be much easier. (Sorry, don't have the time to look at the tests right now)

• Add tests.

@Hahnfeld I added several tests. Because these tests require unified memory to be supported by the underlying system I have added them as a new type of test: check-libomptarget-nvptx-unified (in addition to the check-libomptarget-nvptx one). These tests should only be run on platforms which support unified memory.

• Improve test.