Wykres commitów

230 Commity

Autor SHA1 Wiadomość Data
Evgeny Mankov ed43699a3a Merge pull request #1759 from emankov/master
[HIP] Unify hipError_t (Step 2)
2019-12-30 19:21:09 +03:00
ansurya 651c7a8e27 Fix texture 3D & 2D layered with N components (#1746)
SWDEV-151670: Issue with 3D texture with 4 components
SWDEV-151671: Issue with 2D layered texture with 4 components

Fixed memcpy when memory is allocated with driver API's.

Github issues: #1755
Fixed 3D default case when array type is not set during memory allocation.
2019-12-30 12:41:42 +05:30
Evgeny Mankov 4aaa2336a8 [HIP] Clean-up deprecated HIP error codes
hipErrorMemoryAllocation -> hipErrorOutOfMemory
hipErrorInitializationError -> hipErrorNotInitialized
hipErrorMapBufferObjectFailed -> hipErrorMapFailed
hipErrorInvalidResourceHandle -> hipErrorInvalidHandle
2019-12-23 17:01:35 +03:00
Alex Voicu 150e690a3a Fix late-coming issues. (#1724)
Implementation for hipMemcpyWithStream.
2019-12-23 19:11:24 +05:30
Rahul Garg a369bd4418 Revert - Changes related to hipMemcpyWithStream (#1718)
Reverting #1673, #1697 and #1707.
Support for hipMemcpyWithStream and memcpy optimizations, will be brought in again once issues seen with these are resolved independently.
2019-12-06 09:51:53 +05:30
Rahul Garg 261a04580f Fix hipMemcpyWithStream (#1707)
* Fix hipMemcpyWithStream
* Add a dtest for the same.
2019-12-04 11:56:23 +05:30
Jeff Daily 6df73e1f12 fix hipMempcy precondition, test for zero size first (#1697) 2019-11-25 19:03:03 -08:00
Alex Voicu 022ac3cb0a General sync memcpy improvements. Add hipMemcpyWithStream (#1673)
* General sync memcpy improvements. Add `hipMemcpyWithStream`

* Update hip_memory.cpp
2019-11-20 21:36:37 +05:30
Jeff Daily 854147e911 General hipMemset improvements (#1495)
* hipMemset et al can use HSA API directly for synchronous cases

* lock and flush stream in hipMemset, hold lock until complete

* move hipMemset async check to front of conditional

* use hsa_amd_memory_fill for additional sync memset cases

code cleanup/review for all memset calls

* Fix inversion of execution mutating value.

* ihipMemsetSync fall back to kernel if HSA memset fails

* Never fallback, never surrender.

* Allow NULL stream.

* Optimise memset kernel. Remove deadwood.

* Update hip_memory.cpp

* Clean up stream logic in sync memset

* Revert "Clean up stream logic in sync memset"

This reverts commit 6117dedf673367f44cc704192573a117a3d92477.
2019-11-07 13:19:54 +05:30
ansurya dc8f556460 Fixed texture 2D mapping for pitched arrays & 3D Texture read (#1415)
Texture 2D image mapping for pitched arrays:
github issue: Texture Object's Buffer seems to be Misaligned #886
JIRA ticket: SWDEV-199313

SWDEV-151670 : Fixed issue with 3D texture with 4 components
SWDEV-151671 : Issue with 2D layered texture with 4 components
2019-11-07 13:17:46 +05:30
Rahul Garg f556e15361 Add stream 2019-10-31 12:15:56 -04:00
Rahul Garg 0718ba0f00 Fix HIP init calls in hipMemcpy2DFromArray 2019-10-31 12:15:56 -04:00
Rahul Garg 14b870d1ce Add hipMemcpy2DfromArray (#1510)
Adds hipMemcpy2DFromArray and hipMemcpy2DFromArrayAsync equivalent to cudaMemcpy2DFromArray and cudaMemcpy2DFromArrayAsync.
2019-10-25 15:43:33 +05:30
Aryan Salmanpour 93c688a0c9 [hip] add support for implicit kernel argument for multi-grid sync (#1456)
* [hip] add support for implicit kernel argument for multi-grid sync

* modified code for calculating the prev_sum

* change the impCoopArg type to size_t

* add memory clean up

* launch init_gws and main kernels into two separate loops
2019-10-24 17:43:30 +05:30
Vladislav Sytchenko 0200aa3a21 Update the declarations of hipMemsetD8, hipMemsetD8Async, hipMemsetD16, hipMemsetD16Async. These functions are type aware and take in as their third argument the number of elements in the buffer, not the buffer size. Change the name of this argument from sizeBytes to count to align with the above description. 2019-10-15 14:18:42 -04:00
ansurya 96890792b0 Added new Memory API's (#1399)
Added new memory API's hipMemAllocPitch, hipMemAllocHost, hipMemsetD16, hipMemsetD16Async, hipMemsetD8Async
Modified to support all scenarios hipMemcpyParam2DAsync, hipMemcpyParam2D.
2019-10-04 13:36:31 +05:30
satyanveshd 1ae4cbed4d Reimplement hipMemGetInfo (#1447)
Addresses SWDEV-136570. hipMemGetInfo changed to compute free memory based on information from kfd instead of relying on hc::am_tracker.
2019-10-01 12:40:36 +05:30
satyanveshd 842e304a9c Fix checks in hipMemcpy[D/H]to[D/H]() (#1397)
* SWDEV-202043 hipMemcpyDtoD() issue
2019-09-16 08:32:38 +00:00
Rahul Garg d9179cbe70 Fix typo and colon in comment 2019-09-13 03:09:01 +05:30
Rahul Garg 2e9b8e9d0d [HACK] Temporary fix for hipFree for hipManagedMalloc 2019-09-13 02:10:21 +05:30
Rahul Garg d433f6fb58 Revert "Using HSA API for hipMemsetAsync (#1346)" (#1381)
This reverts commit 9bbd09b04f.
2019-09-03 05:13:46 +00:00
Rahul Garg a786728939 Fix memcpy with IPC slowness (#1321)
* Fix memcpy with IPC slowness

* Make early erroneous returns

* Real Clean up

* Real Clean up++
2019-08-23 09:19:18 +00:00
Jatin Chaudhary 9bbd09b04f Using HSA API for hipMemsetAsync (#1346) 2019-08-21 10:00:10 +00:00
Rahul Garg fbc9f7e20a Add hipMemcpy3DAsync (#1320)
* Add hipMemcpy3DAsync

* Fix CI build error

* Move back stream resolution to internal function

* Remove stream redefinition and check
2019-08-16 02:13:16 +00:00
Rahul Garg 569f35a258 Add hipMemcpyParam2DAsync (#1296)
* Add hipMemcpyParam2DAsync

* Add NVCC path changes

* Clean up

* Fix build issue

* Fix else use in both sync and async apis
2019-08-09 11:50:37 +00:00
Jeff Daily f337ae1edb consolidate thread local storage (#915)
* all thread local access now through single struct

* clean up old commented-out code, more use of GET_TLS()

* fewer calls to GET_TLS by passing tls as a funtion argument

* revert unnecessary change to printf

* fix failing tests due to TLS change

* fix merge conflicts in ihipOccupancyMaxActiveBlocksPerMultiprocessor
2019-08-05 09:51:02 +00:00
Rahul Garg 1c49943ac3 Change hipErrorUnknown to hipErrorInvalidValue 2019-07-31 00:28:30 +05:30
Evgeny Mankov 299fbd4842 [HIP] Fix segfault on uninitialized struct members in hipArrayCreate and hipArray3DCreate 2019-07-12 16:38:26 +03:00
Evgeny Mankov f0832fd968 [HIP][HIPIFY] Split HIP_ARRAY_DESCRIPTOR struct to HIP_ARRAY_DESCRIPTOR and HIP_ARRAY3D_DESCRIPTOR
[Reason] To be compatible with CUDA [#1133]

Update HIP code, hipify-clang, tests and docs

[TODO] Add support of the corresponding functions on nvcc fallback path
2019-07-11 14:58:16 +03:00
Jatin Chaudhary fcb0a3d4e2 Adding bounds check before hipMemset (#1190)
* Adding bounds check in ihipMemset

* Adding ihipMemPtrGetInfo to hipMemPtrGetInfo
2019-07-08 11:00:38 +00:00
Anusha Godavarthy Surya 4989452413 Added missing NULL checks and corrected API return values as per validation 2019-06-27 00:19:05 +05:30
Evgeny Mankov 9cb3e9aa5e [HIP][HIPIFY] Make hipMemcpyParam2D coherent with cuMemcpy2D
+ Makes hip_Memcpy2D struct compatible with CUDA_MEMCPY2D struct
+ Add hipMemcpyParam2D support in nvcc fallback path
+ Update hipify-clang, tests and docs accordingly
2019-05-22 18:31:39 +03:00
Rahul Garg e1f3dc0c80 Add fine grained host memory lock support (#1095)
* Add fine grained host memory lock support

* Fix default flag check
2019-05-13 11:48:26 +05:30
Rahul Garg 94769fc8dd Add hipMallocManaged default functional support (#1036)
* Add hipMallocManaged default functional support

* Fix build error

* Add dtest
2019-04-24 16:50:03 +05:30
Jeff Daily cf8fb43e6b In hipFree, synchronize owner of memory (#1018)
* In hipFree, if memory is associated with a device, synchronize that device's streams.

This changes the behavior from synchronizing the currently set TLS device.

* All devices sync in hipFree for _appId=-1 case.

* Revert "All devices sync in hipFree for _appId=-1 case."

This reverts commit 1efb34d6a8426661e45bc5f763422a1147aeac10.

* add HIP_SYNC_FREE env var
2019-04-16 08:35:55 +05:30
Rahul Garg 50d623981e Handle D2D in memcpy2D 2019-03-28 02:21:45 +05:30
Rahul Garg 9b38380c03 Let hipHostMalloc always share/map pinned host ptr 2019-03-26 10:19:13 +05:30
Rahul Garg ad11972f47 Avoid double mapping of devices to hostMalloc buffer 2019-03-25 23:07:05 +05:30
Maneesh Gupta c20d233585 Merge pull request #970 from mangupta/swdev-172995
hipExtMallocWithFlags implementation
2019-03-25 07:46:53 +00:00
Maneesh Gupta 45255ab492 hipExtMallocWithFlags needs hcc workweek 19115 or higher 2019-03-25 11:41:20 +05:30
Maneesh Gupta e44de376f7 hipExtMallocWithFlags implementation
Change-Id: Iee9e119796472200b2933d5e23be60813f33bc75
2019-03-19 11:59:22 +05:30
Rahul Garg af72cde0a1 Add 2D fallback to use copy kernel 2019-03-14 13:03:06 +05:30
Alex Voicu ed48847237 dlopen() fixes (#929)
* Initial attempt to switch over to internally linked state.

* Add missing CMake update.

* hipLaunchKernelGGLImpl must be inline as well. Ensure internal linkage.

* Ensure global retrieval uses internally linked state.

* Hide HC in the implementation. Minimise ADL woes.

* Strange software exists, and must be catered to.

* Use a less spammy mechanism for ensuring internal linkage / non-export.

* Remove leftover internal detail.
2019-03-06 17:31:44 +05:30
Wen-Heng (Jack) Chung 8b7baa0bd9 Address code review comments to use hipDeviceptr_t 2019-03-05 05:51:05 +00:00
Wen-Heng (Jack) Chung 392271f4db Add hipMemsetD32 and hipMemsetD32Async
Add 2 extra memset functions which fills memory with integer-typed data

Also change the parameters of ihipMemset to better explain the semantic
2019-03-04 17:00:33 +00:00
Wilkin Chau 99540373cf Fix hipMemset3D test
Calculate the allocated size based on the width, height and depth.
2019-02-28 22:42:46 +00:00
Evgeny 47625cb8fd fixing HSA_INIT_API cid args 2019-01-16 23:45:44 -06:00
Maneesh Gupta a778f7cdf7 Merge pull request #797 from gargrahul/fixhipPointerGetAttributes
Fixed hipPointerGetAttributes for hostmalloced ptr
2018-12-12 10:16:07 +05:30
Maneesh Gupta 6ce99b066c Merge pull request #608 from gargrahul/add_pinned_2d_sdma_copy
Added support for pinned 2D SDMA copy
2018-12-12 07:44:16 +05:30
Rahul Garg 77fd517e09 Fixed hipPointerGetAttributes for hostmalloced ptr 2018-12-08 01:42:08 +05:30