AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity
6,802 Commits 3 Branches 0 Tags
8c69e2e52eaf5b2d9cea3dce3f18cb002098ea26
Commit Graph

6088 Commits

Author SHA1 Message Date
sdashmiz 8c69e2e52e SWDEV-377862 - wrong format for memcpy1d 
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: Ide6d252083b1812b0ad9cd182b1435e3e59b3aa1


[ROCm/clr commit: e34be4fbee]
 2023-02-14 09:47:04 -05:00
Sourabh Betigeri 3bc1a92b66 SWDEV-375194 - Fixes graph dot file descriptions to include labelId matching CUDA 
Change-Id: Ia0e6b77e38678b13457f86755ed8f3b852c9b7a1


[ROCm/clr commit: 191b468d89]
 2023-02-13 19:43:12 -05:00
Sourabh Betigeri 1544843291 SWDEV-375202 - Fixes print formatting for flags in hipGraphDebugPrint() to match CUDA 
Change-Id: I2d85fc38d2c65bc12534109883fe00802e77e62d


[ROCm/clr commit: 4cbf9a4a16]
 2023-02-09 13:07:57 -05:00
Ranjith Ramakrishnan e9ebd8b072 SWDEV-366831 - File reorg backward compatibility message changed to #error 
Change-Id: I44adb1a4b9bba0bd89edb75be715504ede2e3d08


[ROCm/clr commit: 810c4ac244]
 2023-02-09 12:46:13 -05:00
Satyanvesh Dittakavi 77e78dff33 SWDEV-370552 - Correct the HIPRTC behavior to optimize the ISA only once 
Change-Id: Idaf0ea8294657db3666e9548deb6a9629e0ee718


[ROCm/clr commit: 99556449b6]
 2023-02-09 11:55:41 -05:00
sdashmiz fb107094d6 SWDEV-374378 - correct setparam for memcpy node 
- params should be valid when used for default flag since we support
  unified virtual address space

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I75d40e437b12ee58e72e423bb4818b484ce35b66


[ROCm/clr commit: 5c5b220561]
 2023-02-09 09:39:20 -05:00
Saleel Kudchadker 83ceb0cd10 SWDEV-381633 - Better log 
Change-Id: I5c59d42462fe8a233ef10e26ec67314219aeb167


[ROCm/clr commit: 2e4ce59cf4]
 2023-02-08 10:21:57 -08:00
Jaydeep Patel 670267ec40 SWDEV-380405 - Check for invalid stream for hipMemcpyPeerAsync. 
Change-Id: I6dfccb4d20bb638bd596c071030c68889743d706


[ROCm/clr commit: e940beb459]
 2023-02-08 00:47:35 -05:00
pghafari 24b7348137 SWDEV-377571 - adding scopelock for hipDeviceReset 
Change-Id: I3dd95a40d6abff721a4774f26e99d1162bafdfa1


[ROCm/clr commit: 03f62b90ef]
 2023-02-07 11:29:20 -05:00
Matt Arsenault 5e61121d88 SWDEV-1 - Use proper __builtin_amdgcn_rcph builtin 
Change-Id: Ic0b1b25a5ab7d132cb51425f2a17bc779c762668


[ROCm/clr commit: eacf63673b]
 2023-02-07 10:58:48 -05:00
Matt Arsenault 6035907648 SWDEV-1 - Eliminate rcp_2f16 pseudo intrinsic 
No such wrapper or intrinsic would ever exist because there is no such
underlying instruction.

Change-Id: I6c3f64cd2df2a58edf32037da8f5712868f296ea


[ROCm/clr commit: bebbf3c4ec]
 2023-02-07 10:58:31 -05:00
Ioannis Assiouras d2904ffe47 SWDEV-380687 - Fixed typo in safe atomicAdd for gfx90a 
Change-Id: I87bc6d9e0ef1f564d679c6280c71c4633d3aa619


[ROCm/clr commit: 6704f3d0f0]
 2023-02-07 10:14:34 -05:00
Ioannis Assiouras a6a6f744ee SWDEV-380466 - Fixed typo in unsafe atomicAdd for doubles and gfx940 
Change-Id: I3b9ea4773f94bf6a9b9dfb655062f42aeb3b5eba


[ROCm/clr commit: 24d3d8216b]
 2023-02-07 10:06:17 -05:00
Jaydeep Patel 4b3f70a6b4 SWDEV-380405 - Make hipMemcpyPeer async w.r.t. host and let dst device wait till copy finishes by src device. 
Change-Id: Idbe25404d0c31b93436e190b548f12f327f47a04


[ROCm/clr commit: 5f231c55e3]
 2023-02-07 00:34:45 -05:00
German Andryeyev 395d85ac2e SWDEV-353281 - Switch Graph to operate with stream 
MemPool was designed to use hip::Stream, but graph implementation uses
amd::HostQueue. Hence switch graph to hip::Stream management.

Change-Id: Ia319389de45e4c3c6043d17473279a6f27a13140


[ROCm/clr commit: 0fe03b6494]
 2023-02-06 10:11:55 -05:00
Matthew Arsenault 716ddd1d98 SWDEV-1 - Reapply "SWDEV-1 - Directly call __builtin_amdgcn_fence" 
This reverts commit 7b3cf2e9d7.

Resubmit after https://github.com/ROCmSoftwarePlatform/rocPRIM/commit/ba8a86166ec9c4cf3a40ecde7017feeb1484e7f6

Change-Id: I194559484a5a0d7f4443381fe77218ea73d2def9


[ROCm/clr commit: 8c7391398a]
 2023-02-03 10:54:27 -05:00
Jaydeep Patel 5981bd5c22 SWDEV-379395 - Use getQueue as it handles null stream and flags can be combination. 
Change-Id: If37854a9fc0fc57acb30e652953a7f283831cad8


[ROCm/clr commit: 01c3299425]
 2023-02-01 12:03:42 -05:00
Jaydeep Patel c6da0aa014 SWDEV-377804 - Initial commit to support hipGraphInstantiateFlagAutoFreeOnLaunch 
Change-Id: I7a35becb6c98a6ff70264e141317d98be7457a37


[ROCm/clr commit: f9e27bcdd4]
 2023-02-01 11:51:39 -05:00
Matt Arsenault c77979a365 SWDEV-1 - Use proper builtins for math intrinsics 
Change-Id: If0e420332c718abe0e9f6316c467b6b29b80b183


[ROCm/clr commit: 14477c24d0]
 2023-02-01 07:19:45 -05:00
sdashmiz a081bdd157 SWDEV-361623 - correct remove edge behaviour 
- remove node dependency before checking parents
- reduce edge level acording to new value of node

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: Id4bff1684f7e0b42beeebc4d2e009bfdb507fb5f


[ROCm/clr commit: 34d087da93]
 2023-01-31 12:48:59 -05:00
Rahul Garg b02b0b13f0 SWDEV-379678 - Remove catch build option 
Change-Id: Idedf5e61da1e6c7fbdd9a65e57220875a3b24112


[ROCm/clr commit: 118e7144f8]
 2023-01-30 11:04:03 -05:00
sdashmiz aaf82fb91b SWDEV-374368 - dependency list should not have duplicate 
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I67e1c8203f10916b98408e2004e6c64e96b1933b


[ROCm/clr commit: b251c8fdb2]
 2023-01-30 10:28:35 -05:00
sdashmiz 483f22536f SWDEV-374389 - correct update dependency behaviour 
- nodes should belong to graph
- num of passed dependency cant be larger than graph nodes

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: Ia6f2283546bc44edee705b0483bfe506b7b1177a


[ROCm/clr commit: 23edcd3e23]
 2023-01-30 10:05:52 -05:00
Ajay GunaShekar 01fe2ea09e SWDEV-372757 - Don't destroy null queue in MT 
This reverts commit 3afbe3a155.

Reason for revert: patch does not fix all stream hangs. So another patch was merged which fixes all issues.

Change-Id: I332d1ea29c23747b46b7667fe3e34e0ceefd2b23


[ROCm/clr commit: 1f4176062b]
 2023-01-25 23:41:16 -05:00
sdashmiz e72db4a957 SWDEV-379151 - correct error status 
- if the stream is null ptr and there is capture ongoing return an error

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: Iff33e4aebc253fb7bb56daf5c455722d726c6705


[ROCm/clr commit: a4dd362a1d]
 2023-01-25 15:06:38 -05:00
Rakesh Roy e8b7717abb SWDEV-375004 - Fix rocprim test failure 
- For !__HIPCC_RTC__ case, operator<<(std::ostream& os, const hip_bfloat16& bf16) calls itself in endless recursion
- Convert hip_bfloat16 to float to fix this

Change-Id: I252a656817550caf43c587cebf461474f12b8c0c


[ROCm/clr commit: 8b00f56577]
 2023-01-25 00:01:02 -05:00
Jaydeep Patel 43565f0496 SWDEV-379125 - Check only if there is attr set for graph kernel node. 
Change-Id: I8768b33ad27e75eb753d99d682edf60b31b240df


[ROCm/clr commit: 70bfc6145b]
 2023-01-24 09:51:12 +00:00
Satyanvesh Dittakavi d96114fd22 SWDEV-372396 - Address the errors in hiprtc headers with -Werror 
Change-Id: Ied553216dfc2ff8b5c72de617cb47b6ca8f52063


[ROCm/clr commit: 0ccd41050b]
 2023-01-23 17:25:42 +00:00
Maneesh Gupta 7b3cf2e9d7 SWDEV-378321 - Revert "SWDEV-1 - Directly call __builtin_amdgcn_fence" 
This reverts commit 1de57c06b3.

Reason for revert: Root cause for regression SWDEV-378321

Change-Id: If3116c5ad1a23db1840e5fc019d2fe997b37d229


[ROCm/clr commit: 5b39c346ad]
 2023-01-19 13:01:27 -05:00
German e95245aaf2 SWDEV-377991 - Rmove Liquidflash logic 
Change-Id: I1c2169c1a8c7449c59cc15074af039c48fc1ef31


[ROCm/clr commit: b685930bee]
 2023-01-19 09:58:07 -05:00
pghafari 5fa035dbd4 SWDEV-369551 - invalid stream check in hipMemcpy 
Change-Id: I6762af93cd566a38f7d04994b87a76e744b19043


[ROCm/clr commit: 76b0b4d635]
 2023-01-18 20:02:28 -05:00
pghafari de904d8716 SWDEV-374381 - updated error check 
for hipGraphAddMemcpyNodeFrom/ToSymbol

Change-Id: I4de063a5a9307148bf4da0e8a2edc1d3b01a8b9a


[ROCm/clr commit: d491ee7385]
 2023-01-18 19:59:32 -05:00
Anusha GodavarthySurya f681854310 SWDEV-376407 - Track manual nodes added during capture 
Change-Id: Ic44f842c303eed0eff08b6ee5922976d96a52c71


[ROCm/clr commit: d422ad0ec1]
 2023-01-18 03:27:41 -05:00
Anusha GodavarthySurya e333f99dec SWDEV-373173 - Handle kernel compilation WGP/cumode in occupancy calculation 
Change-Id: Id0225683a578c3275a35f8f807a57b6379746392


[ROCm/clr commit: ce8f441c14]
 2023-01-18 03:27:30 -05:00
Anusha GodavarthySurya a7a887ab05 SWDEV-373173 - Correct max VGPRs and VGPR Granularity in Occupancy calculation 
Change-Id: I36010f3991f809c61f0c4388b8daa8d3e9b8cc9d


[ROCm/clr commit: 27c5189bca]
 2023-01-18 03:27:21 -05:00
Maneesh Gupta a2a134968c SWDEV-1 - Use VERSION file instead of hipvars.pm for versioning 
Change-Id: I9974aec388c9a89f4ed79a974dea99ec1e8257b8


[ROCm/clr commit: e0516a6984]
 2023-01-18 02:39:12 -05:00
pghafari 5959057299 SWDEV-374361, SWDEV-374362, SWDEV-374364, SWDEV-374365 - error checking 
Change-Id: I0b729976ed794239fa54e1ad2965133b319d2f5d


[ROCm/clr commit: a9cbff0a49]
 2023-01-17 23:10:22 -05:00
sdashmiz 11f97823cd SWDEV-374374 - memset memory mapping should not change 
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I4cc3b1ac5a59fe5c9ae060f4d5d189dc63ca450c


[ROCm/clr commit: 52bf09aac6]
 2023-01-17 14:41:24 -05:00
German Andryeyev 36d240ff89 SWDEV-355281 - Use offset from FindMemObj 
Change-Id: I53ce2c1ef2e3a580b8b12312b693c2476f0885d2


[ROCm/clr commit: 39db0ec297]
 2023-01-17 10:01:49 -05:00
pghafari 829bbe4933 SWDEV-369570 - MaxSharedMemoryPerBloc check 
Change-Id: I74d216a5d9316758bb611eb241ba85222194ce8b


[ROCm/clr commit: 2e60edbe0e]
 2023-01-17 07:52:15 -05:00
pghafari 7261e8fa39 SWDEV-374380 - hipGraphAddMemcpyNode1D errorCheck 
Change-Id: I50501de139e3898783c1cc9c91c5ed3ed2564f3b


[ROCm/clr commit: 236a725d67]
 2023-01-17 07:49:23 -05:00
Satyanvesh Dittakavi f94a34e128 SWDEV-373231 - Support hip atomics header with HIPRTC 
Change-Id: Ic500dfb66ed5ef0c4b672dfd0595409d0e8caaf5


[ROCm/clr commit: 842b9d0b56]
 2023-01-17 00:31:07 -05:00
sdashmiz 42a08627ef SWDEV-360031 - check for stream capture finish. 
- stream capture should be done before any sync APIs.

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I3d65f67ee68777be71f97f48d460ccaefdd4e1af


[ROCm/clr commit: be966acb0c]
 2023-01-16 13:09:24 -05:00
sdashmiz 4f74c83b94 SWDEV-374390 - add correct error 
- null stream should return error on host function

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I17659ad7af0814091dbf6cb4a2cb2644514551f0


[ROCm/clr commit: 338307ef2c]
 2023-01-16 13:08:23 -05:00
Jaydeep Patel d74f646173 SWDEV-374360 - Handle free for external memory. 
Change-Id: Ibfaf30eef371499a3c1f931a73cce76c83d8f7f3


[ROCm/clr commit: 7eecbdd9db]
 2023-01-16 13:07:01 -05:00
Jaydeep Patel ffcff763d2 SWDEV-375366 - SWDEV-375351 - Handle HtoH case for graph mem cpy impl. 
Change-Id: I5a8c3c3c22db045f714b0443b8d70a8c6b4a8cea


[ROCm/clr commit: fdecc1b351]
 2023-01-16 11:37:57 -05:00
Jaydeep Patel c9fa0aad8b SWDEV-369961 - Add mappings for enum hipGraphInstantiateFlags and types. 
Change-Id: I03fa8221684b08a0c44e12aaf2eb7ce5281be629


[ROCm/clr commit: 8bd45ae5da]
 2023-01-15 12:17:21 +00:00
Ioannis Assiouras 1974ef22cf SWDEV-374403 - Added input sanity checks for hipLaunchCooperativeKernelMultiDevice and hipExtLaunchMultiKernelMultiDevice 
Change-Id: Iebef5ae68439d87549fdf9559de51c6c79c64392


[ROCm/clr commit: 0d2e0d1621]
 2023-01-13 19:48:42 -05:00
Matt Arsenault 967ad97993 SWDEV-1 - Use proper builtin for groupstaticsize 
Remove the now empty header for asm intrinsic wrappers.

Change-Id: I63495dc6ed899b1ea823062f2be2fb064af8d4f9


[ROCm/clr commit: 8615a89f24]
 2023-01-13 13:14:10 -05:00
German Andryeyev 9c811dc241 SWDEV-353281 - Initial support of memalloc in graph 
Add memory allocation support in graph. Current implementation uses
cache from mempool  to hold the allocations which belong to the graph.
Also the resource tracking is disabled at this moment because mempool
operates with hip::Stream objects, but graph has execution with
amd::HostQueue objects.

Change-Id: I54fe3250126d24f5a26ada975f37d429bb4ef17b


[ROCm/clr commit: eef47ca24a]
 2023-01-13 13:06:59 -05:00