Commit Graph

573 Commitit

Tekijä SHA1 Viesti Päivämäärä
Ben Sander b1d3df6484 Deprecate hipMallocHost and hipFreeHost.
These will print compiler warnings if used, so we can weed them out
before removing.

Also add a default flags args for hipHostAlloc, in the C++ functioin
headers.  So you can replace hipMallocHost(&ptr, size( with hipHostAlloc(&ptr, size)


[ROCm/clr commit: 57365eb7a3]
2016-03-19 22:53:59 -05:00
Ben Sander b1fe0120ca Refactor copy code.
-Move staging buffer locks inside the staging buffer code.
-Remove dedicated per-device completion_signal + per-device lock -
instead allocated signal from the per-stream pool.   This elimintes
the lock and allows more concurrency.
-remove switch HIP_DISABLE_BIDIR_MEMCPY


[ROCm/clr commit: e64174f47a]
2016-03-18 03:02:00 -05:00
Ben Sander a0d3c018c0 Refactor staging buffer and sync copies.
- refactor staging buffer to operate on hsa* data structures not
  hc::accelerator.
- use hsa_memory_allocate to allocate staging buffers rather than
  am_alloc.
- Refactor device reset with single member function.  Don't reallocate
  staging buffers on reset.
- Properly track dependencies based on command type.  Add new deps for
  H2D and D2D rather than overloading H2D.


[ROCm/clr commit: 3b45e064f9]
2016-03-17 20:09:10 -05:00
Ben Sander dd90a8ff34 Refactor to isolate staging buffer code.
[ROCm/clr commit: 1b7cc7d921]
2016-03-17 00:20:56 -05:00
Ben Sander 622902d408 Start separaration of staging_buffer.cpp code.
Still #include staging_buffer.cpp into hip_hcc.cpp.
Directed tests compile hip_hcc to static library and use the library.


[ROCm/clr commit: a1879ba59b]
2016-03-16 22:26:49 -05:00
Ben Sander 19a0fb0e0e Merge branch 'privatestaging' of https://github.com/AMDComputeLibraries/HIP-privatestaging into privatestaging
Conflicts:
	src/hip_hcc.cpp
	tests/src/CMakeLists.txt


[ROCm/clr commit: 15a8e8f8a0]
2016-03-14 15:01:26 -05:00
Ben Sander a24528abcb enable DB, comments
[ROCm/clr commit: 0d05517d0a]
2016-03-14 14:40:41 -05:00
Ben Sander d6ad50c2e0 Improve error reporting.
use throw with error class.
fix bug when memcpyDefault resolved to D2D copy.


[ROCm/clr commit: ac272932f6]
2016-03-12 04:02:04 -06:00
Aditya Atluri 6162990f44 Added hipHostRegister for hip with tests and added copyright
[ROCm/clr commit: 3127969d97]
2016-03-08 12:57:22 -06:00
Aditya Atluri cf92bfb8c7 Added hipHostRegister flags
[ROCm/clr commit: ffeba62a74]
2016-03-07 10:52:40 -06:00
Aditya Atluri 8a8836088b Added hipHostRegister feature for CUDA backend and its tests
[ROCm/clr commit: 496c549141]
2016-03-07 03:42:50 -06:00
Ben Sander ad3972fdcd Enhance HIP trace debug functions.
- Control with HIP_DB=mask (env var).  See src/hip_hcc.cpp for mask
  values:
    #define DB_API    0 /* 0x01 - shortcut to enable HIP_TRACE_API on single switch */
    #define DB_SYNC   1 /* 0x02 - trace synchronization pieces */
    #define DB_MEM    2 /* 0x04 - trace memory allocation / deallocation */
    #define DB_COPY1  3 /* 0x08 - trace memory copy commands. . */
    #define DB_SIGNAL 4 /* 0x10 - trace signal pool commands */
- Combine with HIP_TRACE to see debug with API trace.
- Use colors to distinguish different flows of debug.
- Add define COMPILE_DB_TRACE to allow removing all debug at compile-time


[ROCm/clr commit: 9b1b108ea8]
2016-03-06 23:50:52 -06:00
Maneesh Gupta e0e67b0e0c Fix typo in nvcc_detail/hip_runtime_api.h
[ROCm/clr commit: b62040f6fd]
2016-03-07 09:40:15 +05:30
Aditya Atluri d7d689cf32 added feature for hipHostGetFlags for CUDA and HIP
[ROCm/clr commit: 45408db5dc]
2016-03-06 12:17:30 -06:00
Aditya Atluri e5657dec23 corrected hipDeviceGetProperties to hipGetDeviceProperties - not docs
[ROCm/clr commit: 8a21b42943]
2016-03-06 08:31:04 -06:00
Aditya Atluri 387bd2e1a7 Added hipHostAlloc with hipHostAllocMapped flag
[ROCm/clr commit: 411154f93f]
2016-03-05 15:57:56 -06:00
Aditya Atluri 375d90a632 Added hipHostAlloc feature for CUDA
[ROCm/clr commit: 2212d35e2d]
2016-03-05 13:58:56 -06:00
Aditya Atluri 28aae1e6c6 v2 Added canHostMapMemory
[ROCm/clr commit: 6085d94f7b]
2016-03-05 13:15:07 -06:00
Aditya Atluri 8254ed4dfa Revert "Added canMapHostMemory feature"
This reverts commit 7eb8b2cc1d.


[ROCm/clr commit: a8d30da648]
2016-03-05 13:08:57 -06:00
Aditya Atluri 7eb8b2cc1d Added canMapHostMemory feature
[ROCm/clr commit: 8c3777d317]
2016-03-05 13:06:37 -06:00
Aditya Atluri 2f61b0c708 Added canMapHostMemory to hipDeviceProp
[ROCm/clr commit: 1e4d1002a0]
2016-03-05 19:30:29 -06:00
pensun 83250c6a54 resolve conflicts of doc_update
[ROCm/clr commit: cb352a17c3]
2016-02-27 15:08:45 -06:00
Aditya Avinash Atluri a44710cd7a Merge pull request #4 from AMDComputeLibraries/memtracker
hipGetPointerAttrib behavioral changes

[ROCm/clr commit: 9c4819bc29]
2016-02-27 10:51:23 -06:00
Aditya Avinash Atluri e33fedcf9f Added CUDA support for hipPointerGetAttributes
[ROCm/clr commit: a31f878218]
2016-02-26 12:33:55 -06:00
Ben Sander b97c2c02b1 fixes for titan platform
[ROCm/clr commit: 8105bd636f]
2016-02-26 05:25:30 -06:00
Ben Sander e345f23846 Merge branch 'memtracker' into privatestaging
Conflicts:
	include/nvcc_detail/hip_runtime_api.h


[ROCm/clr commit: 7a1b4c3878]
2016-02-26 06:17:05 -06:00
Ben Sander ee153bb572 Merge branch 'privatestaging' of https://github.com/AMDComputeLibraries/HIP-privatestaging into privatestaging
[ROCm/clr commit: 4a6173fe58]
2016-02-26 06:15:09 -06:00
Ben Sander b8b7596d4d Merge branch 'memtracker' into privatestaging
Conflicts:
	src/hip_hcc.cpp


[ROCm/clr commit: af97f5e317]
2016-02-25 19:38:46 -06:00
Evgeny Mankov 8e6e28df60 Attribute hipDeviceAttributeIsMultiGpuBoard for obtaining Device property isMultiGpuBoard is added.
On HIP path property obtaining done through hsa_iterate_agents and counting the devices of HSA_DEVICE_TYPE_GPU type.

P.S.
On multi-boards systems it might be problems with detection what board a GPU plugged into (not tested).


[ROCm/clr commit: 7bb0f17656]
2016-02-25 23:44:39 +03:00
Ben Sander 9e9b4fb547 Fix memcpy for Titan. Add <threads> to common includes
[ROCm/clr commit: 784ebcbc86]
2016-02-22 15:09:23 -06:00
Ben Sander 3b8a545ba7 Merge branch 'memtracker' of https://github.com/AMDComputeLibraries/HIP-privatestaging into memtracker
[ROCm/clr commit: 16b04fc0d3]
2016-02-22 08:33:47 -06:00
gargrahul 9bb7be6891 Update for shared atomics support
[ROCm/clr commit: 14508fd0d6]
2016-02-22 16:21:52 +05:30
Ben Sander 4752321fb1 Track last command to a stream.
Passing simple tests.


[ROCm/clr commit: d5c777268a]
2016-02-20 11:02:07 -06:00
Evgeny Mankov c76791140d Guard #ifdef USE_ROCR_20 is added for ROCR_20 device properties (memoryClockRate, memoryBusWidth)
By default isn't defined.
To add ROCR_20 support HIP have to be compiled as follows: make CXX_DEFINES+=-DUSE_ROCR_20


[ROCm/clr commit: d4b15399f5]
2016-02-19 13:27:03 +03:00
Evgeny Mankov a17733dd80 Device property memoryBusWidth implementation.
+ Device property memoryBusWidth is added to hipDeviceProp_t struct.
+ Device attribute hipDeviceAttributeMemoryBusWidth is added to hipDeviceAttribute_t struct.
+ Tests update.


[ROCm/clr commit: da8169dd89]
2016-02-18 18:15:01 +03:00
Evgeny Mankov a47073f25d Device property memoryClockRate implementation.
+ Device property memoryClockRate is added to hipDeviceProp_t struct.
+ Device attribute hipDeviceAttributeMemoryClockRate is added to hipDeviceAttribute_t struct.
+ Tests update.
+ Rename hipDevAttrConcurrentKernels to hipDeviceAttributeConcurrentKernels.


[ROCm/clr commit: 8aace64dce]
2016-02-18 17:25:28 +03:00
Evgeny Mankov 3faa6fd86c Attribute hipDevAttrConcurrentKernels for obtaining Device property concurrentKernels is added.
[ROCm/clr commit: d4bd94e9a0]
2016-02-18 14:34:18 +03:00
Ben Sander a3ec0ae280 remove extra :
[ROCm/clr commit: 866e64f6e2]
2016-02-18 03:05:53 -06:00
Ben Sander 8ed32daefa Remove HIP-local AM tracker (now in HCC)
[ROCm/clr commit: b08e468c06]
2016-02-17 21:33:32 -06:00
Ben Sander 0e83efe14d Add per-stream pool for hsa_signals.
[ROCm/clr commit: 5d721a2649]
2016-02-16 01:59:13 -06:00
Ben Sander f8f40e07bf Update before checkin to HCC.
Add support for USE_AM_TRACKER=2 (HCC version).
Add AM_ALLOC, AM_FREE indirection to ease swapping AM implementations.


[ROCm/clr commit: 1ed431c0f6]
2016-02-15 21:16:00 -06:00
Ben Sander 93c07bc3d1 Move warpSize to header, have shuffles use default warpsize.
[ROCm/clr commit: bd7e3b83b9]
2016-02-15 05:41:09 -06:00
Ben Sander 84810268c0 Update docs, cleanup
[ROCm/clr commit: 322a3bd9b2]
2016-02-15 05:40:12 -06:00
Ben Sander 89e461988e Step1 in staging buffer copy.
- use StagingBuffer class for copies.
- refactor g_device to use array rather than vector.
   (keeps pointers from moving).


[ROCm/clr commit: 90af462b85]
2016-02-12 18:24:08 -06:00
Ben Sander 5978d5f372 Query tracked memory sizes.
Support more accurate hipMemGetInfo.  Add test to hipPointerAttrib.


[ROCm/clr commit: f464cedcf4]
2016-02-12 18:24:08 -06:00
Ben Sander 2089e549eb Tracker improvements
- add API to add / remove user-pointers from the tracker.
- test for thread-safety with MultiThreadtest_2 - rapid
  insertions/removal.
- add mutex to provide thread-safety.
- rename tracker interface to "memtracker_..." for consistency.
- add am_memtracker_reset, connect to hipDeviceReset.
-


[ROCm/clr commit: 7216727fba]
2016-02-12 18:24:08 -06:00
Ben Sander fe67be1134 Create address tracker for am_alloc.
Tracks device where memory is allocated, pinned-host or device, and
more.

Uses memory-range-based lookups - so pointers that exist anywhere in

the range of hostPtr + size will find the associated AmPointerInfo.

The insertions and lookups use a self-balancing binary tree and
should support O(logN) lookup speed.


[ROCm/clr commit: 721508cc2f]
2016-02-12 18:24:08 -06:00
Evgeny Mankov 6add51ef8c Fix typo: maxThreadsPerMultiProcessor -> MaxSharedMemoryPerMultiprocessor
Device property MaxSharedMemoryPerMultiprocessor set equal to totalGlobalMem (HIP path).
Reason: MaxSharedMemoryPerMultiprocessor should be as the same as group memory size. Group memory will not be paged out, so, the physical memory size = total shared memory size = group region size. NVCC path remains untouched: CUDA's device property MaxSharedMemoryPerMultiprocessor is reported.

hipify is updated as well.


[ROCm/clr commit: 460b501cbb]
2016-02-12 01:29:20 +03:00
Evgeny Mankov a8b7647f8b BDFID (BusID/DeviceID/FunctionID) support.
Except FunctionID (or DomainID in CUDA) support, because cudaDeviceProp::pciDomainID is not reported by CUDA.


[ROCm/clr commit: 658e9f0484]
2016-02-11 22:26:01 +03:00
Evgeny Mankov 2478fc078f Formatting, no functional changes
[ROCm/clr commit: d9a94191f2]
2016-02-10 17:21:18 +03:00