AI/rocm-systems
2
0
Atdalīts 0
Kods Problēmas Izmaiņu pieprasījumi Darbības Pakotnes Projekti Laidieni Vikivietne Aktivitāte
6 802 Revīzijas 3 Atzari 0 Tagi
e34be4fbeee7f3682bbe3fae0fc4c94a484d0bfc
Revīziju grafs

6802 Revīzijas

This Branch
This Branch
Visi atzari
Autors SHA1 Ziņojums Datums
sdashmiz e34be4fbee SWDEV-377862 - wrong format for memcpy1d 
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: Ide6d252083b1812b0ad9cd182b1435e3e59b3aa1
 2023-02-14 09:47:04 -05:00
Sourabh Betigeri 191b468d89 SWDEV-375194 - Fixes graph dot file descriptions to include labelId matching CUDA 
Change-Id: Ia0e6b77e38678b13457f86755ed8f3b852c9b7a1
 2023-02-13 19:43:12 -05:00
Sourabh Betigeri 4cbf9a4a16 SWDEV-375202 - Fixes print formatting for flags in hipGraphDebugPrint() to match CUDA 
Change-Id: I2d85fc38d2c65bc12534109883fe00802e77e62d
 2023-02-09 13:07:57 -05:00
Ranjith Ramakrishnan 810c4ac244 SWDEV-366831 - File reorg backward compatibility message changed to #error 
Change-Id: I44adb1a4b9bba0bd89edb75be715504ede2e3d08
 2023-02-09 12:46:13 -05:00
Satyanvesh Dittakavi 99556449b6 SWDEV-370552 - Correct the HIPRTC behavior to optimize the ISA only once 
Change-Id: Idaf0ea8294657db3666e9548deb6a9629e0ee718
 2023-02-09 11:55:41 -05:00
sdashmiz 5c5b220561 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
 2023-02-09 09:39:20 -05:00
Saleel Kudchadker 2e4ce59cf4 SWDEV-381633 - Better log 
Change-Id: I5c59d42462fe8a233ef10e26ec67314219aeb167
 2023-02-08 10:21:57 -08:00
Jaydeep Patel e940beb459 SWDEV-380405 - Check for invalid stream for hipMemcpyPeerAsync. 
Change-Id: I6dfccb4d20bb638bd596c071030c68889743d706
 2023-02-08 00:47:35 -05:00
pghafari 03f62b90ef SWDEV-377571 - adding scopelock for hipDeviceReset 
Change-Id: I3dd95a40d6abff721a4774f26e99d1162bafdfa1
 2023-02-07 11:29:20 -05:00
Matt Arsenault eacf63673b SWDEV-1 - Use proper __builtin_amdgcn_rcph builtin 
Change-Id: Ic0b1b25a5ab7d132cb51425f2a17bc779c762668
 2023-02-07 10:58:48 -05:00
Matt Arsenault bebbf3c4ec 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
 2023-02-07 10:58:31 -05:00
Ioannis Assiouras 6704f3d0f0 SWDEV-380687 - Fixed typo in safe atomicAdd for gfx90a 
Change-Id: I87bc6d9e0ef1f564d679c6280c71c4633d3aa619
 2023-02-07 10:14:34 -05:00
Ioannis Assiouras 24d3d8216b SWDEV-380466 - Fixed typo in unsafe atomicAdd for doubles and gfx940 
Change-Id: I3b9ea4773f94bf6a9b9dfb655062f42aeb3b5eba
 2023-02-07 10:06:17 -05:00
Jaydeep Patel 5f231c55e3 SWDEV-380405 - Make hipMemcpyPeer async w.r.t. host and let dst device wait till copy finishes by src device. 
Change-Id: Idbe25404d0c31b93436e190b548f12f327f47a04
 2023-02-07 00:34:45 -05:00
German Andryeyev 0fe03b6494 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
 2023-02-06 10:11:55 -05:00
Matthew Arsenault 8c7391398a SWDEV-1 - Reapply "SWDEV-1 - Directly call __builtin_amdgcn_fence" 
This reverts commit 5b39c346ad.

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

Change-Id: I194559484a5a0d7f4443381fe77218ea73d2def9
 2023-02-03 10:54:27 -05:00
Jaydeep Patel 01c3299425 SWDEV-379395 - Use getQueue as it handles null stream and flags can be combination. 
Change-Id: If37854a9fc0fc57acb30e652953a7f283831cad8
 2023-02-01 12:03:42 -05:00
Jaydeep Patel f9e27bcdd4 SWDEV-377804 - Initial commit to support hipGraphInstantiateFlagAutoFreeOnLaunch 
Change-Id: I7a35becb6c98a6ff70264e141317d98be7457a37
 2023-02-01 11:51:39 -05:00
Matt Arsenault 14477c24d0 SWDEV-1 - Use proper builtins for math intrinsics 
Change-Id: If0e420332c718abe0e9f6316c467b6b29b80b183
 2023-02-01 07:19:45 -05:00
sdashmiz 34d087da93 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
 2023-01-31 12:48:59 -05:00
Rahul Garg 118e7144f8 SWDEV-379678 - Remove catch build option 
Change-Id: Idedf5e61da1e6c7fbdd9a65e57220875a3b24112
 2023-01-30 11:04:03 -05:00
sdashmiz b251c8fdb2 SWDEV-374368 - dependency list should not have duplicate 
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I67e1c8203f10916b98408e2004e6c64e96b1933b
 2023-01-30 10:28:35 -05:00
sdashmiz 23edcd3e23 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
 2023-01-30 10:05:52 -05:00
Ajay GunaShekar 1f4176062b SWDEV-372757 - Don't destroy null queue in MT 
This reverts commit f39655c0c8.

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

Change-Id: I332d1ea29c23747b46b7667fe3e34e0ceefd2b23
 2023-01-25 23:41:16 -05:00
sdashmiz a4dd362a1d 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
 2023-01-25 15:06:38 -05:00
Rakesh Roy 8b00f56577 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
 2023-01-25 00:01:02 -05:00
Jaydeep Patel 70bfc6145b SWDEV-379125 - Check only if there is attr set for graph kernel node. 
Change-Id: I8768b33ad27e75eb753d99d682edf60b31b240df
 2023-01-24 09:51:12 +00:00
Satyanvesh Dittakavi 0ccd41050b SWDEV-372396 - Address the errors in hiprtc headers with -Werror 
Change-Id: Ied553216dfc2ff8b5c72de617cb47b6ca8f52063
 2023-01-23 17:25:42 +00:00
Maneesh Gupta 5b39c346ad SWDEV-378321 - Revert "SWDEV-1 - Directly call __builtin_amdgcn_fence" 
This reverts commit 3e54bad5b3.

Reason for revert: Root cause for regression SWDEV-378321

Change-Id: If3116c5ad1a23db1840e5fc019d2fe997b37d229
 2023-01-19 13:01:27 -05:00
German b685930bee SWDEV-377991 - Rmove Liquidflash logic 
Change-Id: I1c2169c1a8c7449c59cc15074af039c48fc1ef31
 2023-01-19 09:58:07 -05:00
pghafari 76b0b4d635 SWDEV-369551 - invalid stream check in hipMemcpy 
Change-Id: I6762af93cd566a38f7d04994b87a76e744b19043
 2023-01-18 20:02:28 -05:00
pghafari d491ee7385 SWDEV-374381 - updated error check 
for hipGraphAddMemcpyNodeFrom/ToSymbol

Change-Id: I4de063a5a9307148bf4da0e8a2edc1d3b01a8b9a
 2023-01-18 19:59:32 -05:00
Anusha GodavarthySurya d422ad0ec1 SWDEV-376407 - Track manual nodes added during capture 
Change-Id: Ic44f842c303eed0eff08b6ee5922976d96a52c71
 2023-01-18 03:27:41 -05:00
Anusha GodavarthySurya ce8f441c14 SWDEV-373173 - Handle kernel compilation WGP/cumode in occupancy calculation 
Change-Id: Id0225683a578c3275a35f8f807a57b6379746392
 2023-01-18 03:27:30 -05:00
Anusha GodavarthySurya 27c5189bca SWDEV-373173 - Correct max VGPRs and VGPR Granularity in Occupancy calculation 
Change-Id: I36010f3991f809c61f0c4388b8daa8d3e9b8cc9d
 2023-01-18 03:27:21 -05:00
Maneesh Gupta e0516a6984 SWDEV-1 - Use VERSION file instead of hipvars.pm for versioning 
Change-Id: I9974aec388c9a89f4ed79a974dea99ec1e8257b8
 2023-01-18 02:39:12 -05:00
pghafari a9cbff0a49 SWDEV-374361, SWDEV-374362, SWDEV-374364, SWDEV-374365 - error checking 
Change-Id: I0b729976ed794239fa54e1ad2965133b319d2f5d
 2023-01-17 23:10:22 -05:00
sdashmiz 52bf09aac6 SWDEV-374374 - memset memory mapping should not change 
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I4cc3b1ac5a59fe5c9ae060f4d5d189dc63ca450c
 2023-01-17 14:41:24 -05:00
German Andryeyev 39db0ec297 SWDEV-355281 - Use offset from FindMemObj 
Change-Id: I53ce2c1ef2e3a580b8b12312b693c2476f0885d2
 2023-01-17 10:01:49 -05:00
pghafari 2e60edbe0e SWDEV-369570 - MaxSharedMemoryPerBloc check 
Change-Id: I74d216a5d9316758bb611eb241ba85222194ce8b
 2023-01-17 07:52:15 -05:00
pghafari 236a725d67 SWDEV-374380 - hipGraphAddMemcpyNode1D errorCheck 
Change-Id: I50501de139e3898783c1cc9c91c5ed3ed2564f3b
 2023-01-17 07:49:23 -05:00
Satyanvesh Dittakavi 842b9d0b56 SWDEV-373231 - Support hip atomics header with HIPRTC 
Change-Id: Ic500dfb66ed5ef0c4b672dfd0595409d0e8caaf5
 2023-01-17 00:31:07 -05:00
sdashmiz be966acb0c 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
 2023-01-16 13:09:24 -05:00
sdashmiz 338307ef2c SWDEV-374390 - add correct error 
- null stream should return error on host function

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I17659ad7af0814091dbf6cb4a2cb2644514551f0
 2023-01-16 13:08:23 -05:00
Jaydeep Patel 7eecbdd9db SWDEV-374360 - Handle free for external memory. 
Change-Id: Ibfaf30eef371499a3c1f931a73cce76c83d8f7f3
 2023-01-16 13:07:01 -05:00
Jaydeep Patel fdecc1b351 SWDEV-375366 - SWDEV-375351 - Handle HtoH case for graph mem cpy impl. 
Change-Id: I5a8c3c3c22db045f714b0443b8d70a8c6b4a8cea
 2023-01-16 11:37:57 -05:00
Jaydeep Patel 8bd45ae5da SWDEV-369961 - Add mappings for enum hipGraphInstantiateFlags and types. 
Change-Id: I03fa8221684b08a0c44e12aaf2eb7ce5281be629
 2023-01-15 12:17:21 +00:00
Ioannis Assiouras 0d2e0d1621 SWDEV-374403 - Added input sanity checks for hipLaunchCooperativeKernelMultiDevice and hipExtLaunchMultiKernelMultiDevice 
Change-Id: Iebef5ae68439d87549fdf9559de51c6c79c64392
 2023-01-13 19:48:42 -05:00
Matt Arsenault 8615a89f24 SWDEV-1 - Use proper builtin for groupstaticsize 
Remove the now empty header for asm intrinsic wrappers.

Change-Id: I63495dc6ed899b1ea823062f2be2fb064af8d4f9
 2023-01-13 13:14:10 -05:00
German Andryeyev eef47ca24a 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
 2023-01-13 13:06:59 -05:00