Improve debug print messages.

- Remove "call-to-call" for hipStreamCreate and hipEventCreate.
  These now call an internal functions rather than calling through
  hipStreamCreateWithFalgs and hipEventCreateWithFlags.

- Add HIP_INIT_API for more functions so they trace correctly.

- Use stream#DEVICE.STREAMID in debug messages via new specialization in
  tace_helper.


[ROCm/hip commit: 7934cf620d]

projects/hip/include/hcc_detail/hip_hcc.h

@@ -341,7 +341,7 @@ typedef uint64_t SeqNum_t ;
 
     //-- Non-racy accessors:
     // These functions access fields set at initialization time and are non-racy (so do not acquire mutex)
-    ihipDevice_t *       getDevice() const;
+    ihipDevice_t *              getDevice() const;
     StreamMutex &               mutex() {return _mutex;};
 
     //---
@@ -375,9 +375,20 @@ private:
     std::deque<ihipSignal_t>    _signalPool;   // Pool of signals for use by this stream.
 
     StreamMutex                  _mutex;
+
+    friend std::ostream& operator<<(std::ostream& os, const ihipStream_t& s);
 };
 
 
+inline std::ostream& operator<<(std::ostream& os, const ihipStream_t& s)
+{
+    os << "stream#";
+    os << s._device_index;
+    os << '.';
+    os << s._id;
+    return os;
+}
+
 
 //----
 // Internal event structure:

projects/hip/include/hcc_detail/hip_runtime_api.h

@@ -443,10 +443,7 @@ hipError_t hipStreamCreateWithFlags(hipStream_t *stream, unsigned int flags);
  * @see hipStreamDestroy
  *
  */
-static inline hipError_t hipStreamCreate(hipStream_t *stream)
-{
-    return hipStreamCreateWithFlags(stream, hipStreamDefault);
-}
+hipError_t hipStreamCreate(hipStream_t *stream);
 
 
 /**
@@ -544,13 +541,10 @@ hipError_t hipEventCreateWithFlags(hipEvent_t* event, unsigned flags);
 /**
  *  Create an event
  *
- *  @param[in] event Creates an event
+ * @param[in,out] event Returns the newly created event.
  *
  */
-static inline hipError_t hipEventCreate(hipEvent_t* event)
-{
-    return hipEventCreateWithFlags(event, 0);
-}
+hipError_t hipEventCreate(hipEvent_t* event);
 
 
 /**

projects/hip/include/hcc_detail/trace_helper.h

@@ -44,16 +44,38 @@ std::string ToHexString(T v)
 
 
 //---
-// Template overloads for ToString to handle various types:
-// Note these use C++11 variadic templates
+// Template overloads for ToString to handle specific types
+
+// This is the default which works for most types:
 template <typename T>
-std::string ToString(T v) {
+std::string ToString(T v) 
+{
     std::ostringstream ss;
     ss << v;
     return ss.str();
 };
 
 
+//  hipEvent_t specialization. TODO - maybe add an event ID for debug?
+template <>
+std::string ToString(hipEvent_t v) 
+{
+    return ToString(&v);
+};
+
+
+
+//  hipStream_t
+template <>
+std::string ToString(hipStream_t v) 
+{
+    std::ostringstream ss;
+    ss << *v;
+
+    return ss.str();
+};
+
+//  hipMemcpyKind specialization
 template <>
 std::string ToString(hipMemcpyKind v) {
     switch(v) {