AI/rocm-systems
2
0
Bifurcation 0
Code Tickets Demandes d'ajout Actions Paquets Projets Publications Wiki Activité
1 400 Révisions 3 Branches 0 Étiquette
a6ac4e7097df73bf5a847a536ca69a0f54e686b6
Graphe des révisions

1400 Révisions

Cette branche
Cette branche
Toutes les branches
Auteur SHA1 Message Date
scchan a6ac4e7097 calls isfinite,isinf,isnan from the std namespace on the host 
Change-Id: Ica2370075b89713eecfd96102e2f4e0ab9961ce4
 2017-02-14 11:52:09 -05:00
Evgeny Mankov 353f5b238a [HIPIFY] Add more unhandled refs for opaque pointers. 
CUfunc_st -> hipFunction_t *
CUctx_st -> hipCtx_t *
CUmod_st -> hipModule_t *
 2017-02-14 17:34:10 +03:00
Evgeny Mankov ce2daa53fb [HIPIFY] Fix warnings on unhandled refs. 
CUstream_st -> hipStream_t *
CUevent_st -> hipEvent_t * (fixed, was -> hipEvent_t)

Warning:
[HIPIFY] warning: the following reference is not handled: 'CUstream_st' [struct var ptr]. (thrust/examples/cuda/async_reduce.cu)
 2017-02-14 17:23:09 +03:00
Evgeny Mankov afe6073d3f [HIPIFY] Narrow cudaKernelCallExpr matcher. 
[Explanation]
Narrow cudaKernelCallExpr matcher to isExpansionInMainFile() - to the file being actually converted (without system includes).
But as for Thrust, there are *.inl files, which are nevertheless being expanded in main file.
These files contain cuda kernel launches, which are not converted successfully by hipify, for instance:
thrust/system/cuda/detail/detail/launch_closure.inl:98:23.

[ToDo]
File a bug on cudaKernelCallExpr matcher for conversion error (thrust\examples\cuda\<any cu example>).
 2017-02-14 16:28:05 +03:00
Evgeny Mankov e5c601df45 [HIPIFY] Minor refactoring 
+ insertHipHeaders function is added to Cuda2Hip class;
+ fix replacement ending for hip header.
 2017-02-13 18:13:49 +03:00
Maneesh Gupta 57f7f4660a Generate package config for cmake 
Change-Id: I73a0b21361bbe29e095dc515bdc70588ec722d57
 2017-02-13 16:01:50 +05:30
Paul f1f01ceb63 Set hip version correctly 2017-02-13 15:57:10 +05:30
Aditya Atluri 1cffe49775 v3: added free for ihipModuleSymbol_t structures inside tracker 
Change-Id: Ib8041a05312c08cbdf2d4fee5e7cbae17df6efff
 2017-02-10 13:42:10 -06:00
Aditya Atluri 554fe191e1 v2: Fixed hipModule memory management 
1. Changed test to assert for same hipFunction values
2. Added better memory management for hipModule

Change-Id: I10d7aef13c215a2211e262f3c79017f26a17d9a7
 2017-02-10 13:32:13 -06:00
Evgeny Mankov 25d4de9e95 [HIPIFY] 'CHANGED LOC', 'TOTAL LOC' and 'CODE CHANGED (in lines)' are added to statistics. 2017-02-10 22:02:41 +03:00
Maneesh Gupta 3967ece8dd Build libhip_hcc.so as a fat binary 
Change-Id: Ie4f334d8f9576edf5df0f917f72083d7842eb193
 2017-02-10 10:32:04 +05:30
Aditya Atluri 307a21c946 fixed shared memory usage implementation for hipDynamicShared2.cpp test 
Change-Id: I34c72cb905f78de0f37e94174382e8be5c532028
 2017-02-09 18:17:42 -06:00
Aditya Atluri d6b5c02862 added new dynamic shared mem test 
Change-Id: Ic2a12fc9bc5b67e85f1e6c6008f10c7c66388377
 2017-02-09 18:07:48 -06:00
Aditya Atluri 6110fcdfb6 fixed hipFunction memory management 
Change-Id: I7ebb323419bcd220ebd6466a8eb38e7bfdb1520a
 2017-02-09 17:22:55 -06:00
Maneesh Gupta 7f62529b1f Update hip_hcc packaging script 
Change-Id: I2dc96545c472942bcf2bc8a2e88735f3ba26d1e1
 2017-02-09 20:28:28 +05:30
Maneesh Gupta 6659eb8d16 Disable failing directed tests 
- hipTestDeviceSymbol
- hipTestConstant
- hipTestMallocKernel

Change-Id: Ibfe9fc0b8a59882f1de64b42e18777a7bd56ee97
 2017-02-09 14:48:22 +05:30
Maneesh Gupta 71363810dd Fix arch specific hcc extra libs env var 
Change-Id: I7429da2f1cb98750d6a9601e7e5bde844a098487
 2017-02-09 14:38:39 +05:30
Aditya Atluri c6c90f6f5a added architecture specific macros 
1. Added __HIP_ARCH_GFXNUM__
2. Usage, -D__HIP_ARCH_GFX803__=1

Change-Id: I68b3a85d62cfab3a45d2b7a70cb3518ab2565236
 2017-02-08 19:45:32 -06:00
pensun f2bfc8a405 Add pseudo code example for hip_bugs.md 
Change-Id: Ia2af8e6165faeb3fbb81428e20d4dc5b19b2fa9e
 2017-02-08 16:20:01 -06:00
Aditya Atluri 400d0d4f78 fixed hipcc for new compiler flags 
Change-Id: I49ec059be20ff26b7482c84d91ab7a43826c6a8d
 2017-02-08 14:06:01 -06:00
Aditya Atluri 26b5e57cfd include arch specific ir on fallback path 
Change-Id: Ib04996aae2c21eb73ef2a9f6305915e0caccd704
 2017-02-08 12:19:06 -06:00
Aditya Atluri 60ec83c683 Fixed Hawaii link issues 
1. Split hip_ir.ll to hip_hc.ll and hip_hc_gfx803.ll
 a. hip_hc.ll contains arch generic ir implementations
 b. hip_hc_gfx803.ll contains gfx803 (fiji, polaris) specific ir
2. HIPCC can now parse --amdgpu-target=*.
 a. Usage: hipcc --amdgpu-target=gfx803 --amdgpu-target=gfx701
 b. TODO: Convert to --amdgpu-target=gfx803,gfx701
3. With LLC in HCC able to generate native f16 isa, removed inline half asm math ops
4. Fixed threadfence and threadfence_block to use functions in rocdl

Change-Id: Ic9a9e3e04139b0d75d2c2a263c030ca77adc1019
 2017-02-08 12:04:05 -06:00
Evgeny Mankov cc8a7f3d3e [HIPIFY] Readme.md update. 2017-02-08 19:50:05 +03:00
Evgeny Mankov b3098091fd [HIPIFY] Readme.md update. 
+ Launching example.
+ Delimiter "--" is strongly recommended.
+ "-x cuda" option is mentioned as well.
 2017-02-08 19:43:32 +03:00
Aditya Atluri be6d01ebc3 Fixed HIP for C runtime 
1. fixed constant memory test
2. added runtime gcc/g++ test
3. added fix for runtime with gcc/g++

Change-Id: Ie14dad6633411b188bdeea044e140b6d5beabe16
 2017-02-07 13:15:36 -06:00
Aditya Atluri 6d8007e28c fixed vector types for c 
Change-Id: I2330e976060f9a038929cd51be07ae2ee98e81ce
 2017-02-07 11:09:54 -06:00
Rahul Garg 73b56e690f Command scripts for latency measurements 
Change-Id: I8c28765a09fb0358447367939de524b12699a317
 2017-02-07 15:03:46 +05:30
Maneesh Gupta 0bef861f1a Disable broken test: hipC 
Change-Id: I061aa125bbdc9f14bc870266ab0735593c861903
 2017-02-06 13:00:50 +05:30
Aditya Atluri f009f3533d changed __global__ attribute 
1. Moved around tests and added them to HIT

Change-Id: I5d75280c42a5af852670ebabc7305ee56721ec7b
 2017-02-03 10:53:36 -06:00
Aditya Atluri 0d4e6ae60a fixed symbol memcpy issue 
Change-Id: I89d7401be51d194bcbf771020ba66e3d3b6a18f8
 2017-02-01 17:54:59 -06:00
Aditya Atluri a5d017e406 removed host math functions from math_functions.h 
Change-Id: I90d8784e2d6b58c6fade9f0fa12c0db3ee417d3e
 2017-01-27 17:38:43 -06:00
Aditya Atluri eee520def9 changed device code tests to work not to work as one 
Change-Id: I0eec1eab19dda3b703bc3a0d778a6bbb2802a412
 2017-01-27 09:20:14 -06:00
Aditya Atluri a8c51b822b fixed arch compiler flag 
Change-Id: I19f7a0ea513d6b8059f0c10cd0e7a5ead510e246
 2017-01-27 08:51:48 -06:00
Aditya Atluri 537dc7597b fixed threadfence ir 
Change-Id: Ia3afb54bdb50864e678d849608d72a3c321edba1
 2017-01-27 08:42:26 -06:00
Aditya Atluri 33d89d06aa fixed kernel only compilation for module api 
Change-Id: I567992fa9b87125318edba41fd82d2f7bc5504a1
 2017-01-27 08:41:42 -06:00
Rahul Garg 125983eed9 hipMallocArray fixes for NV path 
Change-Id: I1ca43e6bc0cd405998888005c20dfb1ea57003d5
 2017-01-27 14:32:08 +05:30
pensun f18efd5d15 more fix on hipmallocarray on NV path 
Change-Id: I890a36cab10c101f4a112bc4567f765b318d486c
 2017-01-26 12:30:52 -06:00
pensun 42f22162e0 fix missing semicolon on NV path 
Change-Id: I3cfecb7bd534578a1f5a07ca9397092dcf01db07
 2017-01-26 11:28:15 -06:00
Ben Sander 1a24178c78 Add HIP_FAIL_SOC. 
Fail sub-optimal-copies rather than perform them slowly.
SOC occur on async copy of unpinned memory, or P2P copy between GPUs
that are not peers.
 2017-01-25 21:53:17 -06:00
Ben Sander 2f7a8ec39c Read HCC_OPT_FLUSH and optimize dispatch accordingly. 
If HCC is in this mode, we can use less aggressive flushes in some
cases.
 2017-01-25 21:50:52 -06:00
Aditya Atluri 33a3ebad27 added device functions header to hip_runtime.h 
Change-Id: I635931e1fbe4b7f0f64f3a126c0f1badcd6d234b
 2017-01-25 08:14:30 -06:00
pensun a8fcbb2584 Add more hip_bug.md entry, regarding hang after hipLaunchKernel 
Change-Id: I5800cb627179ec0e913cd36d332fb8c2994ab71e
 2017-01-24 22:43:25 -06:00
pensun 24d9827688 Initial commit on hip_bugs markdown doc 
Change-Id: I5a6915337b8664cfed9eaee9443c6e4406348574
 2017-01-24 22:30:36 -06:00
Ben Sander 73ef364c74 Add debug tip to docs 2017-01-23 22:34:41 -06:00
Ben Sander 27447f10c2 Show dynamic shared mem usage not static. 2017-01-23 22:34:41 -06:00
Ben Sander bc809460f5 Move core env var processing to env.cpp 2017-01-23 22:34:41 -06:00
Ben Sander 4de3df746c Add debug tips to docs 2017-01-23 22:34:41 -06:00
Ben Sander fe24996326 Add debug tips to docs 2017-01-23 22:34:41 -06:00
Ben Sander d19c4767b7 Log error with ihipLogError. Cleans up CXL trace display. 2017-01-23 22:34:41 -06:00
Ben Sander df74158d1c Add HIP_IGNORE_HCC_VERSION. 
Ignores strict checking of HCC and HIP version.
Can be useful when developing new HCC code.
 2017-01-23 22:34:41 -06:00