[Openmp-commits] [PATCH] D133463: [OpenMP] Device Time Profile

[Johannes Doerfert via Phabricator via Openmp-commits]

jdoerfert added a comment.

I left some initial comments, I think this is not in a "working" state, it looks like parts are missing, beyond the LLVM changes.



================
Comment at: openmp/libomptarget/DeviceRTL/src/Kernel.cpp:25
 
+// CLOCKINIT();
+extern DeviceTimeProfileTy omptarget_device_time_profile;
----------------
Leftover


================
Comment at: openmp/libomptarget/DeviceRTL/src/Kernel.cpp:45
+    return OMP_ARRAY_SIZE;
+  }
+}
----------------
Turn this around.
```
if (Idx >= OMP_ARRAY_SIZE)
  return OMP_ARRAY_SIZE
...
```

Where does OMP_ARRAY_SIZE and OMP_THREADS come from? Shouldn't those be read from the struct the runtime initializes?


================
Comment at: openmp/libomptarget/DeviceRTL/src/Kernel.cpp:50
+  if (Idx < OMP_ARRAY_SIZE) {
+#pragma omp parallel
+    {
----------------
Don't use parallel here. This is the device runtime.
Depending on where this is called, all threads will execute it anyway.
Just use `mapping::getThreadIdInBlock()` to get the thread id and `mapping::getBlockId()` to get the block id.
Same above.


================
Comment at: openmp/libomptarget/DeviceRTL/src/Kernel.cpp:120
 
+  CLOCKIN();
   if (IsSPMD) {
----------------
What is this?


================
Comment at: openmp/libomptarget/DeviceRTL/src/Kernel.cpp:170
   FunctionTracingRAII();
+  CLOCKOUT();
   const bool IsSPMD = Mode & OMP_TGT_EXEC_MODE_SPMD;
----------------
And this?


CHANGES SINCE LAST ACTION
  https://reviews.llvm.org/D133463/new/

https://reviews.llvm.org/D133463