6425 Υποβολές

Συγγραφέας SHA1 Μήνυμα Ημερομηνία
Rahul Garg d859888adb Fix 2D and 3D memset (#1987)
[ROCm/hip commit: 6c65fc04d1]
2020-04-06 15:35:59 +05:30
ansurya f59ed79983 Initial support for bfloat16 (#1980)
[ROCm/hip commit: 770e76e752]
2020-04-06 15:35:43 +05:30
Yaxun (Sam) Liu c636253a71 Fix ambiguity of fma for _Float16 for libc++ (#1976)
libc++ defines fma as template function for auto promotion of mixed-type
arguments. libc++ does not handle _Float16 as _Float16 is not a supported
type by C++ standard. As such, it is unlikely we can commit our fix for
_Float16 to libc++ trunk.

Therefore we handle _Float16 with a template specialization of
__numeric_type in HIP headers.

Change-Id: If01960a657ebf1a7a67463cdcf66fab7458dff3c

[ROCm/hip commit: 4af2106d10]
2020-04-06 15:35:18 +05:30
Sameer Sahasrabuddhe 49184503f8 SWDEV-227201: Introduce tests for printf on hostcall
Tests that check POSIX specifiers with a single thread:
 - hipPrintfSpecifiers.cpp     : all conversion specifiers
 - hipPrintfFlags.cpp          : common flags that modify conversions
 - hipPrintfAltForms.cpp       : alternate forms ('#')
 - hipPrintfStar.cpp           : additional arguments ('*')
 - hipPrintfWidthPrecision.cpp : floating point details

Tests that check functionality on top of hostcall
 - hipPrintfBasic.cpp       : divergent calls, series of calls, return value, etc
 - hipPrintfManyWaves.cpp   : many waves printing together
 - hipPrintfManyDevices.cpp : many waves on many devices

Change-Id: I35e069f4c542f896999239996dc89eda0faad7b8


[ROCm/hip commit: db3e17d552]
2020-04-06 00:49:34 -04:00
eshcherb 4f4701c323 adding HIP_PROF version macro (#2000)
[ROCm/hip commit: e4a1c49e72]
2020-04-04 13:10:28 -07:00
Christophe Paquot d4df8f2042 Default HostMalloc to uncached memory
Change-Id: I72e19c7f7820a77fd5afc09f09cfea9acd0b8e84


[ROCm/hip commit: fa5a9b3810]
2020-04-03 19:19:33 -04:00
Saleel Kudchadker 5f95e68b90 Merge "Wake up commandQueue before returning" into amd-master-next
[ROCm/hip commit: f1a9a4f22a]
2020-04-03 18:27:03 -04:00
Michael LIAO 5bdf843642 [vdi] Add hipFreeHost
Change-Id: I8a5b7ff3f0ab4f5674efd6723c18808ad6ef33f5


[ROCm/hip commit: 9e619430f4]
2020-04-03 16:34:28 -04:00
German Andryeyev 73c507c3c6 Merge "SWDEV-184709 - support hipLaunchCooperativeKernel()" into amd-master-next
[ROCm/hip commit: 6f19e77f69]
2020-04-03 16:23:05 -04:00
Vladislav Sytchenko 0d6c4fe470 Take into an account the number of channels...
when querying the element size of an array.

Change-Id: Id57d3374b14d80a59230ec8286704f2fbabb0fae


[ROCm/hip commit: 5f14ae1161]
2020-04-03 15:43:18 -04:00
German Andryeyev e55b231bb1 SWDEV-184709 - support hipLaunchCooperativeKernel()
- Add validation checks for cooperative launch to pass Cuda test

Change-Id: Ie296f0c3f113909d9a357879db3b2a833ab314c5


[ROCm/hip commit: 7820018037]
2020-04-03 15:18:21 -04:00
Michael Hong Bin Liao 8c63e2536d Merge "Fix size type in __hipRegisterVar" into amd-master-next
[ROCm/hip commit: fad9e39597]
2020-04-03 15:12:21 -04:00
Saleel Kudchadker f3bdfe2baa Wake up commandQueue before returning
Change-Id: I87eb5a22c81a9cb807474a960b5987d5fb6c2b86


[ROCm/hip commit: bc8d6ac97c]
2020-04-03 10:23:36 -07:00
Aaron En Ye Shi 1ed78d9869 Merge "Fix path for hip-clang when using hipcc (#1961)" into amd-master-next
[ROCm/hip commit: a2104f6c94]
2020-04-03 12:28:50 -04:00
Michael LIAO 9759e31dac Fix size type in __hipRegisterVar
Change-Id: I6b667600ae8f133583b768ab963318882b84179f


[ROCm/hip commit: 00bb2ce04a]
2020-04-03 10:51:58 -04:00
Michael Hong Bin Liao 7083af4690 Merge "[hip] Clean up unnecessary casting." into amd-master-next
[ROCm/hip commit: dbedce9e54]
2020-04-03 10:50:40 -04:00
German Andryeyev 97926390dc SWDEV-184709 - support hipLaunchCooperativeKernel()
- Enable cooperative tests for single and multiple devices

Change-Id: I54b6713f578b6b5e670f117b17469c0091028c99


[ROCm/hip commit: 0d69b79622]
2020-04-02 12:55:05 -04:00
Michael LIAO 4b28e810f9 [hip] Clean up unnecessary casting.
Change-Id: I64b08aaef5c67ffb49330c9c605611f1fbd3f5a2


[ROCm/hip commit: 79d8f7e47e]
2020-04-02 12:46:15 -04:00
Paul Fultz II 17f9e2358b Fix path for hip-clang when using hipcc (#1961)
* Fix path for hip-clang when using hipcc

* Fix typo

* Update regex

Change-Id: I31bbee2e70d58b89191f970f5c6ae7e1c8b40900


[ROCm/hip commit: c8bcf28320]
2020-04-02 12:09:31 -04:00
Paul Fultz II 8aff07fe80 Add missing flags for hip::device target on hip-clang (#1230)
This adds the missing compilation flags to hip::device so it can compile with hip-clang compiler.

Change-Id: Ie2b30ea606bfca385a0e84ae03ee0a8d828ad16a


[ROCm/hip commit: 2c246dabfe]
2020-04-02 12:09:03 -04:00
Saleel Kudchadker 124e126d85 Merge "OpenCL2.2 Header changes" into amd-master-next
[ROCm/hip commit: 68d013d030]
2020-04-02 02:46:49 -04:00
Vladislav Sytchenko 1ce5ee07d2 Add entry points for hipTexObject*() API
Even though the runtime and driver texture object API is one to one, the structs used by these APIs are not. See hipResourceDesc vs HIP_RESOURCE_DESC differences.

These differences are not trivial and most likely won't be able to handled by hipify, so we need new API entry points.

Change-Id: Id4bcb1ad0ae15378dbdb5a2ed07e5ea30f320082


[ROCm/hip commit: aea688b79c]
2020-04-01 14:51:51 -04:00
eshcherb 40a323cad4 adding hipApiString (#1989)
[ROCm/hip commit: 8c4e8d6484]
2020-04-01 12:10:33 +05:30
ansurya 1344399fa2 tex1Dfetch behaviour for different address mode and filter mode (#1772)
Fixes github issue: #1754

- When ResourceDesc::resType is hipResourceTypeLinear ignore address mode and filter mode.
- When textureDesc::normalizedCoords is set to zero, AddressModeWrap and AddressModeMirror won't be supported and will be switched to AddressModeClamp.

[ROCm/hip commit: 50ef250a3b]
2020-04-01 12:10:17 +05:30
Saleel Kudchadker 1bd55da10a Merge "Cleanup stream from hip:Event class." into amd-master-next
[ROCm/hip commit: 34a3ed9c1b]
2020-03-31 20:14:48 -04:00
Vladislav Sytchenko deacfcd063 (SWDEV-229354)
This patch is a workaround to support user pitch for hipMemcpy{2D/3D}.

Historically OpenCL didn't support pitch with clEnqueueFillBuffer(), so neither did we in VDI. Adding it now will be slightly nontrivial, since the fill kernel and runtime in many places will need to be modified.

As a temporary workaround for cases when pitch > width, we can just enqueue a fill for each row separately. This implementation is slow, but it satisfies the correctness criteria.

Change-Id: Idfeca349288b51d6ff84a7cf001fb63c6a66818a


[ROCm/hip commit: 77223a8eca]
2020-03-31 18:12:56 -04:00
Reshabh Sharma 264082869c Output file name should not change flags picked for compiler (#1938)
Fixes SWDEV-207362,

The output file name should not contribute to picking up the right flags for the compiler. This fix solves issues when the output has conflicting extensions which confuses hipcc to treat them as the source files and add the required flags for them.

PS: Output file refers to the file followed by -o

Change-Id: I1095966c11143ad73e81fabc35b4e9de5d3afada
Example: hipcc test.o -o test.hip will add the flags for .hip compilation ignoring the fact that it is an output file


[ROCm/hip commit: bb62f8553f]
2020-03-31 16:13:43 -04:00
Saleel Kudchadker 2f1617e21a Cleanup stream from hip:Event class.
Change-Id: I98de07d33bb7fea8f5e2d32b288c15f10ce58902


[ROCm/hip commit: e4ea1f60bb]
2020-03-31 11:22:00 -07:00
Michael Hong Bin Liao e89abbb63b Merge "[hipcc] Remove the previous workaround." into amd-master-next
[ROCm/hip commit: 5b96f38c1e]
2020-03-31 10:05:56 -04:00
Michael LIAO 5e5f4e23d5 [vdi] Fix hipGetSymbol{Address|Size}
- Use symbol value as the qeury key. Compared to the symbol name, the
  symbol value is more robust as developers may use unqualified or
  qualified identifiers. It also removes the mangling and/or demangling
  requirement for the runtime API.

Change-Id: I9d4259f3842612c7cc98551269fc2092d8b5c19e


[ROCm/hip commit: b72196613a]
2020-03-31 00:26:53 -04:00
Christophe Paquot 24cd888b3a Do not retry to allocate when OOM. Shouldn't be needed since we idle on Free.
SWDEV-229214

Change-Id: I183006f409388e3c7981f2569649d01d6378be46


[ROCm/hip commit: 94a7ef6ed1]
2020-03-30 12:49:48 -07:00
kjayapra-amd ce461331d3 SWDEV-216213 - Use different static & dynamic module maps for faster lookup.
Change-Id: Ia605e76a411ad5be04046b9d61f1ac111d49bb4a


[ROCm/hip commit: b081912962]
2020-03-30 14:28:07 -04:00
Anusha Godavarthy Surya 03a8b95eb4 Merge "Update Enable/Disable peers to match cuda behaviour" into amd-master-next
[ROCm/hip commit: caa083357a]
2020-03-30 13:46:42 -04:00
agodavar d7db609b92 Update Enable/Disable peers to match cuda behaviour
Change-Id: I67194ccf77a0019368579ff7d95b7790fcf228f3


[ROCm/hip commit: bdb3a4b393]
2020-03-30 12:49:16 -04:00
Tao Sang 1254e84b7b Merge "enable HCC printf when using hip-clang" into amd-master-next
[ROCm/hip commit: 1697cfc790]
2020-03-30 11:37:23 -04:00
Michael LIAO 9a2e585762 [hipcc] Remove the previous workaround.
Change-Id: Ie1c2aa21689bc343c10fd4b699546cf500ced5b2


[ROCm/hip commit: f1566f6dc5]
2020-03-30 10:18:47 -04:00
Michael LIAO cbac75a003 [vdi] Fix calculation of MaxWaves
- Consider the case where `usedVGPRs` is zero.
- This fixes [SWDEV-228537](http://ontrack-internal.amd.com/browse/SWDEV-228537)

Change-Id: I8675311f5fe24fb59c5d45bada122afefb55b128


[ROCm/hip commit: 3a690e960f]
2020-03-30 09:10:16 -04:00
Sarbojit2019 2292982475 Fix for segfault seen in hipMemcpyDtoD (#1964)
* Fixes SWDEV-227444.

[ROCm/hip commit: eba596c87a]
2020-03-28 17:29:49 +05:30
Maneesh Gupta 0ee021d444 Remove address_space(1) typecast and use __ockl_atomic_add_noret_f32 (#1956)
* Remove address_space(1) typecast for ockl_global_atomic_add_f32
* use __ockl_atomic_add_noret_f32

[ROCm/hip commit: cbc3d1713f]
2020-03-28 17:28:33 +05:30
Sameer Sahasrabuddhe 3e7d08f5ee enable HCC printf when using hip-clang
This is cherry-picked from PR#1947 that was committed to the
github repo. It allows printf to work with hip-clang and HCC
runtime.

Change-Id: I754753250ea1e694cf3441722e2d4c9d25fa75bc


[ROCm/hip commit: 9a0c5d0653]
2020-03-28 00:18:21 -04:00
Saleel Kudchadker 35e9fa08be Merge "Check event status before notify" into amd-master-next
[ROCm/hip commit: 5a8add03a5]
2020-03-27 20:19:40 -04:00
Vladislav Sytchenko 511f2d03a2 (SWDEV-228794)
Adjust the origin of the copy if the user passes a pointer that wasn't allocated by the runtime.

Change-Id: I0aeb20195ed730857a461a53f537626ec2573fd1


[ROCm/hip commit: 6ed73f50f7]
2020-03-27 16:33:16 -04:00
Vladislav Sytchenko d3b9203359 (SWDEV-228794)
Add hipMallocHost()

Change-Id: Ia3c7c5ca94b39fe30f3a51d1b60782d3472259ff


[ROCm/hip commit: fd7a8f0367]
2020-03-27 15:57:48 -04:00
Vladislav Sytchenko 7a867cd98f (SWDEV-228782)
The only requirment from hipMallocPitch() is that the returned pitch is aligned to the HW image pitch alignment. There is no restriction on the size of the allocation, since the memory might not be used for images.

Change-Id: I97438e5fe4012ca4721b14b85f514dbac803c17c


[ROCm/hip commit: a91b82f00e]
2020-03-27 15:52:17 -04:00
Tao Sang aa94d9c654 Merge "Support hipFuncGetAttributes with hip-clang+Hcc RT" into amd-master-next
[ROCm/hip commit: af7d131749]
2020-03-27 14:36:41 -04:00
Saleel Kudchadker 573d676ef0 Check event status before notify
Change-Id: I68f6bbbf236e49b859be2d5afbe0c8282fe15dd3


[ROCm/hip commit: 178fc17846]
2020-03-27 11:32:46 -07:00
satyanveshd 3eea799a53 [dtests] Added few Negative tests (#1735)
[ROCm/hip commit: 351d39e6aa]
2020-03-27 14:10:12 +05:30
Siu Chi Chan e58a0d06f7 don't expose symbols from code_object_bundle (#1971)
Change-Id: I56479485aad42c3d517fe6d9055be1cd846eeb00

[ROCm/hip commit: 43abf84f54]
2020-03-27 14:09:07 +05:30
Sarbojit2019 1909e436cf Fix few memory leaks in HIP (#1969)
[ROCm/hip commit: f1b028b93e]
2020-03-27 14:08:30 +05:30
Jatin Chaudhary 32a9b2f2bb Packaging changes to bring them to sync with generated ones (#1918)
[ROCm/hip commit: 3805d7c5f2]
2020-03-27 14:08:12 +05:30