Grafico dei commit

1914 Commit

Autore SHA1 Messaggio Data
Siu Chi Chan bec0ff43e0 Merge remote-tracking branch 'origin/master' into HEAD
[ROCm/clr commit: cf98d0e37d]
2017-10-27 01:18:28 -04:00
Ben Sander b0aa15ee5f Merge pull request #198 from AlexVlx/feature_support_globals_for_module_api
Feature support globals for module api

[ROCm/clr commit: 8a64feef61]
2017-10-27 01:53:34 +02:00
Ben Sander 1058f074f3 Merge pull request #218 from ChrisKitching/nodiscard
Add [[nodiscard]] attribute to hipError_t in C++17 mode

[ROCm/clr commit: 9713dbb6f6]
2017-10-26 22:48:54 +02:00
Ben Sander 7ce81b1573 Merge pull request #223 from bensander/2x_bidir
Use 2X for bidir memory bandwidth calc

[ROCm/clr commit: f76bcf4045]
2017-10-26 21:49:06 +02:00
Evgeny Mankov 2371dda65e Merge pull request #213 from ChrisKitching/simplify
[HIPIFY] Simplify (and accelerate) hipification of CUDA type identifiers

[ROCm/clr commit: fbdedd9196]
2017-10-26 19:17:01 +03:00
Kent Knox f32e20abc7 Update container to newer cuda driver and sdk 9.0
[ROCm/clr commit: f81d6d5bb3]
2017-10-25 16:14:32 -05:00
Siu Chi Chan d97b4b55cc add HC_FEATURE_PRINTF around the printf buffer definition
[ROCm/clr commit: 085d63099c]
2017-10-25 12:00:02 -04:00
Chris Kitching 31920efae1 Don't use now-defunct cmake variable in lit test config
[ROCm/clr commit: 8d91579dcf]
2017-10-24 20:52:51 +01:00
Chris Kitching 5ee04d16c5 Refactor cudaCall to prefer early return to deep nesting
Sorry for the invasive refactor, but this was making reasoning
about this function more difficult.


[ROCm/clr commit: b24b33ee2e]
2017-10-24 20:38:49 +01:00
Chris Kitching 4d23fb75e9 Split the giant lookup table into 3 smaller ones
Instead of having a single, enormous LUT for all CUDA names, let's
have separate ones for different types of entity. We often know
that we're looking at a typename, or a function name, or a macro
name - so we can be more efficient (and resilient to name
collisions) by having smaller lookup tables for each of those
classes of entity).

Here we start that off by having three LUTs:
- Header names
- Type names
- Everything else

Future work could usefully split "everything else" into:
- enum values
- macro names
- function names
- everything else

It's worth noting that the "needs new matcher" todos I delete here
were actually resolved with the previous commit. It no longer
naively searches for things that start with "cu*" - it will find
exactly those things that are present in our lookup tables.


[ROCm/clr commit: 9da456b315]
2017-10-24 20:38:49 +01:00
Chris Kitching a00a5e3101 One matcher for type expressions to rule them all
Previously, there were different AST matchers for each
language construct that contains a type reference, and custom
logic to perform the transformation within each of those
structures.

Since the transformation in all such cases was only replacing
CUDA types with hip ones, we can instead use an AST matcher
that finds and updates the type references directly.
This simplifies the program considerably, and it won't fail
when it finds a language feature (or complicated type expression)
that nobody wrote custom logic for yet.


[ROCm/clr commit: 93c9b3ca34]
2017-10-24 20:38:49 +01:00
Chris Kitching 1a53e259b3 Make control flow less insane
`while(false)` is certainly a bold choice.


[ROCm/clr commit: 3b2d8029fb]
2017-10-24 20:38:49 +01:00
Chris Kitching 02d3a34917 Use a cmake glob for collecting hipify sources
Should make breaking this monstrosity into multiple files a bit
easier...


[ROCm/clr commit: 9a4778f435]
2017-10-24 20:38:48 +01:00
Chris Kitching bbac7c8b82 Move giant lookup table into another translation unit
Also, rewrote it as a constant variable instead of a function
that imperatively fills a map. It's shorter, faster the compile,
and (depending on how badly the compiler screws it up) maybe
faster to run.

And, of course, it starts breaking up that giant .cpp file.


[ROCm/clr commit: 803d3ffd9c]
2017-10-24 20:38:25 +01:00
Evgeny Mankov 5a2bdb98f0 Merge pull request #230 from emankov/master
[HIPIFY][fix] cmake: NO_DEFAULT_PATH is strongly needed in find_package for LLVM

[ROCm/clr commit: 4e78738267]
2017-10-24 20:44:56 +03:00
Evgeny Mankov a12402fc9a Merge branch 'master' into tests
[ROCm/clr commit: 712a1da073]
2017-10-24 20:03:51 +03:00
Evgeny Mankov a6f0e70cea Merge pull request #227 from ChrisKitching/clang-silly
Tweak some version numbers in clang version compatibility checks

[ROCm/clr commit: 7b307b21d1]
2017-10-24 17:16:07 +03:00
Evgeny Mankov 59c0a69751 [HIPIFY][fix] cmake: NO_DEFAULT_PATH is strongly needed in find_package for LLVM
Otherwise LLVM will be searched in system folders.


[ROCm/clr commit: 4a0228cedb]
2017-10-24 16:35:10 +03:00
Rahul Garg 28c72504b6 Example showing globals use with module APIs
[ROCm/clr commit: 10b1b58505]
2017-10-24 18:12:25 +05:30
Evgeny Mankov 10cd9f9eb9 [HIPIFY][fix] cmake: do not build hipify-clang if not asked
+ warn "hipify-clang will not be built" if HIPIFY_CLANG_LLVM_DIR is not specified.
+ fix typo in previous commit .


[ROCm/clr commit: 7c51376bbf]
2017-10-24 14:16:05 +03:00
emankov d9a550d09c [HIPIFY] cmake: simplify build
[ROCm/clr commit: 559d9a68aa]
2017-10-24 10:51:11 +03:00
Siu Chi Chan f3b4833413 printf support for module API
[ROCm/clr commit: 4ed3e9e15f]
2017-10-24 00:55:41 -04:00
Chris Kitching 482541bb8c Tweak some version numbers in clang version compatibility checks
Apparently a couple of those APIs changed in clang 5, not 4.

Drat.


[ROCm/clr commit: aa288fcf1f]
2017-10-24 01:45:23 +01:00
Siu Chi Chan 9997997fe5 add an unit test for printf
[ROCm/clr commit: f1992c5b11]
2017-10-23 18:31:13 -04:00
Siu Chi Chan 1e4a2a7dd0 add a printf device function
[ROCm/clr commit: 36c2795dd4]
2017-10-23 18:30:08 -04:00
Siu Chi Chan 5f24981c61 replace __hcc_workweek__ with HC_FEATURE_PRINTF flag
[ROCm/clr commit: da7c37947c]
2017-10-23 18:30:08 -04:00
Evgeny Mankov 27df5ecc82 [HIPIFY] cmake: fix standalone build
[ROCm/clr commit: cdd849541f]
2017-10-23 21:16:13 +03:00
Rahul Garg dc6c43772c Use 2X for bidir p2p memory bandwidth calc
[ROCm/clr commit: 4090c82936]
2017-10-23 21:57:20 +05:30
Chris Kitching 26be50b082 Add concurentKernels.cu to the testsuite
[ROCm/clr commit: 2c65d0da37]
2017-10-23 13:39:37 +01:00
Chris Kitching 8fffd03350 Add the CUDA samples include dir to the path for tests
Means we get to easily steal CUDA examples for tests


[ROCm/clr commit: 64d5f07050]
2017-10-23 13:39:37 +01:00
Chris Kitching 7547f796d0 Add cudaRegister.cu lit test
[ROCm/clr commit: 2437b31939]
2017-10-23 13:39:37 +01:00
Chris Kitching 8fd3b3b1cd Add square.cu to lit testsuite
[ROCm/clr commit: ead79e5bf4]
2017-10-23 13:39:37 +01:00
Chris Kitching 8b311424a6 Introduce a test runner script to simplify invocation
... And to use a standard, highly amusing trick for making
coloured output work.


[ROCm/clr commit: c99dcbba8d]
2017-10-23 13:39:37 +01:00
Chris Kitching 4e80782cda Adapt lit test for the hipLaunchKernelGGL changes from before...
[ROCm/clr commit: 5912f465bd]
2017-10-23 13:39:37 +01:00
Chris Kitching 0f6c153774 Migrate lit test to using FileCheck, so failures are readable
It seems the test is already broken, but look how awesome the
error message is now:

/home/chris/HIP/tests/hipify-clang/axpy.cu:31:12: error: expected string not found in input
 // CHECK: hipLaunchKernel(HIP_KERNEL_NAME(axpy), dim3(1), dim3(kDataLen), 0, 0, a, device_x, device_y);
           ^
<stdin>:31:2: note: scanning from here
 //
 ^
<stdin>:33:2: note: possible intended match here
 hipLaunchKernelGGL(axpy, dim3(1), dim3(kDataLen), 0, 0, a, device_x, device_y);
 ^


[ROCm/clr commit: 74fd64d5c1]
2017-10-23 13:39:37 +01:00
Chris Kitching d92c43bd21 Look for FileCheck for running lit tests, too
Use of grep in `lit` RUN lines is deprecated:
https://llvm.org/docs/TestingGuide.html#writing-new-regression-tests

Using grep leads to really unhelpful failure output (it literally
just says "the test failed"). FileCheck is much more helpful, and
distributed with LLVM on most distros anyway, so this extra
dependency shouldn't prove problematic.


[ROCm/clr commit: 3868036ea7]
2017-10-23 13:39:36 +01:00
Chris Kitching baabd2755e Propagate the CUDA toolkit directory into the lit tests
Allows the tests to actually run... :D


[ROCm/clr commit: 9747578d09]
2017-10-23 13:39:36 +01:00
Chris Kitching 49e43a7fc1 Remove dependency on nonexistent lit target
This just... isn't a target. Maybe you were trying to do a file
dependency? But that's handled implicitly by the find_program
call anyway, so all this is doing is generating a cmake warning


[ROCm/clr commit: e9d259699c]
2017-10-23 13:39:36 +01:00
Maneesh Gupta 941682c701 Make elfio headers private
Change-Id: I3ba174bb46e84a75380207d93a0da6fe3703689e


[ROCm/clr commit: 281c543bc3]
2017-10-23 10:24:36 +05:30
Ben Sander eee60317fa Use 2X for bidir memory bandwidth calc
[ROCm/clr commit: 21689e8710]
2017-10-21 07:47:32 -05:00
Evgeny Mankov 255a770364 Merge pull request #219 from ChrisKitching/newClang
[HIPIFY] Migrate hipify-clang to a newer version of clang

[ROCm/clr commit: 9dec691dbf]
2017-10-21 11:59:46 +03:00
Ben Sander 33b46b8ed9 Merge pull request #214 from scchan/reinit_printf_buffer
hipDeviceReset(): make sure to reinitialize the printf buffer in hcc RT

[ROCm/clr commit: 326651875d]
2017-10-20 11:04:37 -05:00
Chris Kitching d0feaee4f4 Avoid a double-free of HipifyPPCallbacks instance
This bug was present all along, but something changed in the order
of de-initialisation performed by llvm that makes it actually
crash now.

The constructor of HipifyPPCallbacks gives:
```
std::unique_ptr<HipifyPPCallbacks>(this)
```
to the LLVM Preprocessor instance. The Preprocessor instance
subsequently frees the HipifyPPCallbacks, which is then freed
again when we leave the stack frame at line 4340.

So: let's leak the HipifyPPCallbacks onto the heap, and leave
the LLVM Preprocessor object responsible for tidying it up.


[ROCm/clr commit: 893ee6d6ca]
2017-10-19 17:09:32 +01:00
Chris Kitching 64a26027b7 Be agnostic to the new getReplacements() API
See comment


[ROCm/clr commit: a06fe54ee8]
2017-10-19 17:08:55 +01:00
Chris Kitching 5e09ca10b7 Use inline initialisers to set default field values
A trivial cleanup that helps in a moment..


[ROCm/clr commit: 9d83348ac7]
2017-10-19 17:08:55 +01:00
Chris Kitching fd9bc0d0d8 Omit now-removed Filename string arg from handleBeginSource
[ROCm/clr commit: abd34b2125]
2017-10-19 17:08:55 +01:00
Chris Kitching a21737be39 Cope with Replacements now having llvm::Error returns
[ROCm/clr commit: 885573e1e2]
2017-10-19 17:08:54 +01:00
Chris Kitching db2b03b879 Cope with clang 4.0's rename of getNumArgs()
Sorry, this one I couldn't do in a perfectly elegant way ;)


[ROCm/clr commit: 9ccb966e2b]
2017-10-19 17:08:54 +01:00
Chris Kitching ae40d4bdea Remove unnecessary call to Retain() on a smart pointer
The Preprocessor smart pointer is held by the CompilerInstance,
and therefore its reference count cannot reach zero until the
CompilerInstance itself is destroyed.

If the CompilerInstance is destroyed, you have more to worry about
than just the preprocessor being deallocated!

Newer versions of the LLVM/Clang API migrated to using
std::shared_ptr, so there is no `Retain()` function (by that
name, anyway). Eliminating this redundant use is a neat and
backward-compatible way to become compatible with newer versions
of the LLVM/Clang API.


[ROCm/clr commit: 22a5e2330d]
2017-10-19 17:08:54 +01:00
Chris Kitching 3f284c0927 Refer to clang::StringLiteral explicitly
Newer versions of llvm/clang mean there is both an
llvm::StringLiteral and a clang::StringLiteral. Since we're
dumping both namespaces wholesale into the global namespace with
`using` declarations, this creates a name collision, which must be
resolved.

This change is backwards-compatible, and fixes a problem you
encounter when using newer versions of the llvm/clang API.


[ROCm/clr commit: 73984ed809]
2017-10-19 17:08:54 +01:00