[Openmp-commits] [PATCH] D14254: [OpenMP] Initial implementation of OpenMP offloading library - libomptarget device RTLs.

[Jonas Hahnfeld via Phabricator via Openmp-commits]

Hahnfeld added inline comments.


================
Comment at: libomptarget/deviceRTLs/nvptx/src/debug.cu:21-22
+
+NOINLINE void PrintTaskDescr(omptarget_nvptx_TaskDescr *taskDescr, char *title,
+                             int level) {
+  omp_sched_t sched = taskDescr->GetRuntimeSched();
----------------
This is used only once in `libcall.cu`. Please inline and remove this file entirely


================
Comment at: libomptarget/deviceRTLs/nvptx/src/debug.cu:33-65
+////////////////////////////////////////////////////////////////////////////////
+// debug for compiler (should eventually all vanish)
+////////////////////////////////////////////////////////////////////////////////
+
+EXTERN void __kmpc_print_str(char *title) { PRINT(LD_ALL, " %s\n", title); }
+
+EXTERN void __kmpc_print_title_int(char *title, int data) {
----------------
Never generated by current compilers, remove


================
Comment at: libomptarget/deviceRTLs/nvptx/src/libcall.cu:393-394
+
+  // TODO: should check for the lock to be SET?
+  int ret = atomicCAS(lock, compare, val);
+
----------------
I think the comment is right here: This looks like it is setting the lock if it is unset!


================
Comment at: libomptarget/deviceRTLs/nvptx/src/libcall.cu:401
+
+//for xlf Fotran
+//Fotran, the return is LOGICAL type
----------------
Can you check that these are still used? Especially the `debug` functions?


================
Comment at: libomptarget/deviceRTLs/nvptx/src/omptarget-nvptx.h:100
+};
+extern __device__ __shared__ DataSharingStateTy DataSharingState;
+
----------------
This won't compile with current Clang 5.0.0:
```
error: __shared__ variable 'DataSharingState' cannot be 'extern'
```


Repository:
  rL LLVM

https://reviews.llvm.org/D14254