Update docs for event, review event TODO.

Change-Id: Iec491f9f22df163f01c0af6639fcbe33c81acdcc


[ROCm/hip commit: 61af94a555]

projects/hip/include/hip/hcc_detail/hip_runtime_api.h

@@ -626,8 +626,12 @@ hipError_t hipStreamGetFlags(hipStream_t stream, unsigned int *flags);
  *
  * @param[in,out] event Returns the newly created event.
  * @param[in] flags     Flags to control event behavior.  Valid values are #hipEventDefault, #hipEventBlockingSync, #hipEventDisableTiming, #hipEventInterprocess
- *
- * @warning On HCC platform, flags must be #hipEventDefault.
+ 
+ * #hipEventDefault : Default flag.  The event will use active synchronization and will support timing.  Blocking synchronization provides lowest possible latency at the expense of dedicating a CPU to poll on the eevent.
+ * #hipEventBlockingSync : The event will use blocking synchronization : if hipEventSynchronize is called on this event, the thread will block until the event completes.  This can increase latency for the synchroniation but can result in lower power and more resources for other CPU threads.
+ * #hipEventDisableTiming : Disable recording of timing information.  On ROCM platform, timing information is always recorded and this flag has no performance benefit.
+ 
+ * @warning On HCC platform, hipEventInterprocess support is under development.  Use of this flag will return an error.
  *
  * @returns #hipSuccess, #hipErrorInitializationError, #hipErrorInvalidValue, #hipErrorLaunchFailure, #hipErrorMemoryAllocation
  *

projects/hip/src/hip_event.cpp

@@ -34,15 +34,16 @@ hipError_t ihipEventCreate(hipEvent_t* event, unsigned flags)
 {
     hipError_t e = hipSuccess;
 
-    // TODO - support hipEventDefault, hipEventBlockingSync, hipEventDisableTiming
-    if (flags == 0) {
+    // TODO-IPC - support hipEventInterprocess.
+    unsigned supportedFlags = hipEventDefault | hipEventBlockingSync | hipEventDisableTiming;
+    if ((flags & ~supportedFlags) != 0) {
         ihipEvent_t *eh = new ihipEvent_t();
 
         eh->_state  = hipEventStatusCreated;
         eh->_stream = NULL;
         eh->_flags  = flags;
         eh->_timestamp  = 0;
-        *event = eh; // TODO - allocat the event directly, no copy needed.
+        *event = eh; 
     } else {
         e = hipErrorInvalidValue;
     }
@@ -152,7 +153,6 @@ hipError_t hipEventElapsedTime(float *ms, hipEvent_t start, hipEvent_t stop)
 
             int64_t tickDiff = (stop_eh->_timestamp - start_eh->_timestamp);
 
-            // TODO-move this to a variable saved with each agent.
             uint64_t freqHz;
             hsa_system_get_info(HSA_SYSTEM_INFO_TIMESTAMP_FREQUENCY, &freqHz);
             if (freqHz) {