Commit grafiek

110 Commits

Auteur SHA1 Bericht Datum
Chris Kitching 80120aed22 Introduce a test runner script to simplify invocation
... And to use a standard, highly amusing trick for making
coloured output work.
2017-10-23 13:39:37 +01:00
Chris Kitching 640afd4186 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.
2017-10-23 13:39:36 +01:00
Chris Kitching efa814e381 Propagate the CUDA toolkit directory into the lit tests
Allows the tests to actually run... :D
2017-10-23 13:39:36 +01:00
Chris Kitching 76ccda4205 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
2017-10-23 13:39:36 +01:00
Chris Kitching f5b273fc4f 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.
2017-10-19 17:09:32 +01:00
Chris Kitching 6718519025 Be agnostic to the new getReplacements() API
See comment
2017-10-19 17:08:55 +01:00
Chris Kitching 24fc459f69 Use inline initialisers to set default field values
A trivial cleanup that helps in a moment..
2017-10-19 17:08:55 +01:00
Chris Kitching 7ae6a10c99 Omit now-removed Filename string arg from handleBeginSource 2017-10-19 17:08:55 +01:00
Chris Kitching 7e253365f1 Cope with Replacements now having llvm::Error returns 2017-10-19 17:08:54 +01:00
Chris Kitching 575bedb28c Cope with clang 4.0's rename of getNumArgs()
Sorry, this one I couldn't do in a perfectly elegant way ;)
2017-10-19 17:08:54 +01:00
Chris Kitching 0953a7887d 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.
2017-10-19 17:08:54 +01:00
Chris Kitching 2975d00edc 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.
2017-10-19 17:08:54 +01:00
Chris Kitching 3a2fe40f78 Fix two faulty LLVM version checks
What we actually want to do here is use the StringRef version in
versions newer than 3.8, and the void one in 3.8 and older.

Checking "major-version >= 3 && minor-version >= 9" does not do
what we want. Consider what this will do for version 4.0, for
which minor-version is zero...
2017-10-19 17:08:54 +01:00
Chris Kitching be5a120f90 Use the cache for global variables - not PARENT_SCOPE hacks 2017-10-19 17:08:54 +01:00
Chris Kitching 9a3faec12c Skip lit tests if lit cannot be found 2017-10-19 17:08:54 +01:00
Chris Kitching 764d89dcbe Don't reinvent find_package
PythonInterp is a finder module that ships with cmake. It supports
the conventional interaction with find_package that allows you
to demand success, and particular vesions, without having your
own logic:

https://cmake.org/cmake/help/v3.0/command/find_package.html
2017-10-19 17:08:53 +01:00
Chris Kitching 757b9c3b76 Don't be picky about clang versions 2017-10-19 17:08:53 +01:00
Chris Kitching cb948dc7fd Declare HIPIFY_CLANG_TESTS as a cmake option 2017-10-19 17:08:53 +01:00
Chris Kitching 51ee1bdc2d Use early return to avoid indenting all of CMakeLists.txt 2017-10-19 17:08:53 +01:00
Chris Kitching 517d46c333 Fix broken indentation introduced by previous commit 2017-10-19 17:08:53 +01:00
Chris Kitching 4ba2b2e5b8 Inline a redundant std::vector
Looks like a partial refactor..?
2017-10-19 14:11:25 +01:00
Chris Kitching 742526706b Allow user-provided -std=* argument to override the default
This means any user-provided -std=* argument overrides this one,
allowing use of newer standards.
2017-10-19 14:11:15 +01:00
Evgeny Mankov db7661ccad Merge pull request #210 from ChrisKitching/output-hipify
[HIPIFY] Alter output file behaviour slightly
2017-10-18 21:46:31 +03:00
Chris Kitching 296c5a33ce Unconditionally append .hip as the default output filename 2017-10-18 19:19:18 +01:00
Chris Kitching e4e17a56bb If an output path is given _use it_
Don't append .hip to a user-provided output file...
2017-10-16 17:32:57 +01:00
Chris Kitching 2263cb9f72 Operate on a temp file in-place
This copies to the output after operation, instead of working
_on_ the output. This allows includes to work correctly, while
supporting output paths anywhere on the filesystem.

Fixes #208
Fixes #206
2017-10-16 17:32:56 +01:00
Evgeny Mankov 7ee8e2d51b [HIPIFY][#199][Partial fix] Fix for cudaLaunchKernel transformation
[Synopsis]
If any of kernel arguments is a MACRO its location calculation is wrong (location of its definition is actually calculated).
Thus garbage code is being produced on the place of such a MACRO starting from the end of its actual definition.

[Solution]
Add isMacroBodyExpansion and isMacroArgExpansion checks on kernel arguments.
2017-10-12 20:05:42 +03:00
Evgeny Mankov 87cf949d39 [HIPIFY] cmake changes
+ standalone build support
+ WIN MSVC support
+ disable lit testing by default
+ to enable lit testing -DHIPIFY_CLANG_TESTS=1 should be specified

Tested builds:
+ hipify-clang standalone           {Ubuntu 16.04|Windows 10} {Debug|Release} {x86|x64a} {g++|clang++|MSVC 2017} {llvm 3.8.0|llvm 3.9.0} {-DHIPIFY_CLANG_TESTS=1| }
+ HIP {-DHIPIFY_CLANG_LLVM_DIR=..|} {Ubuntu 16.04}            {Debug|Release} {x86|x64a} {hcc}                   {llvm 3.8.0|llvm 3.9.0} {-DHIPIFY_CLANG_TESTS=1| }
2017-10-09 20:28:20 +03:00
Evgeny Mankov 53ce40ebab [HIPIFY] CUDA RT Textures and Arrays support update 2017-10-05 16:38:36 +03:00
Evgeny Mankov 28a300aed2 [HIPIFY] CUDA RT Textures support update 2017-09-18 17:52:50 +03:00
Evgeny Mankov d614f53ad7 Update Cuda2Hip.cpp 2017-07-28 01:45:24 +03:00
Evgeny Mankov ffcf5d32a0 [HIPIFY] CUDA RT Textures support 2017-07-27 18:39:40 +03:00
Evgeny Mankov 368aebc4bc [HIPIFY] CUDA Stream Callback support 2017-07-26 19:29:50 +03:00
Evgeny Mankov 85e6593a1b [HIPIFY] Finally finished syncing with CUDA 8.0.61 Driver and Runtime… (#105)
[hipify-clang] Finally finished syncing with CUDA 8.0.61 Driver and Runtime API (including missing data types, D3D, OpenGL, VDPAU and EGL interop).

    + All the Modules are supported now:
	1) 4.1 – 4.31 from CUDA 8.0.61 Driver API
	2) CUDA_Driver_API_functions_supported_by_HIP.md updated accordingly
	3) 4.1 – 4.31 from CUDA 8.0.61 Runtime API
	4) CUDA_Runtime_API_functions_supported_by_HIP.md updated accordingly
    + Typos fixing
    + Annotating
2017-07-26 09:38:03 +05:30
Maneesh Gupta 91f82ce541 Update toc in markdown documentation
Change-Id: I6da7053672b306442f3640fff3471efe25593870
2017-06-28 11:32:25 +05:30
Evgeny Mankov d239b1a3fc [HIPIFY] [DOC] Fix typo. 2017-06-23 21:59:24 +03:00
Evgeny Mankov b4a39664f0 [HIPIFY] Sync more CUDA Driver API functions.
+ 4.14. Event Management
+ 4.15. Execution Control

ToDo: 4.16 - 4.31 modules of CUDA Driver API.
2017-06-22 21:53:32 +03:00
Evgeny Mankov 0208fa4e70 [HIPIFY] Sync HIPIFY with HIP by CUDA Driver API functions.
+ 4.12. Unified Addressing
+ 4.13. Stream Management

ToDo: 4.14 - 4.31 modules of CUDA Driver API.
2017-06-14 19:55:55 +03:00
Evgeny Mankov f805bd0c1d [HIPIFY] Initial sync HIPIFY with HIP by CUDA Driver API functions.
+ CUDA_Driver_API_functions_supported_by_HIP.md update.
+ Initial update of HIPIFY with CUDA driver API functions:
  1.Error Handling, 2.Initialization, 3.Version Management, 5-6.Device Management, 7.Primary Context, 8-9.Context, 10.Module Management, 11.Memory Management.
+ Sync HIP functions against CUDA Driver and Runtime API functions.
+ Typo fixes.

ToDo: 12-30 modules of CUDA Driver API.
2017-06-13 20:25:11 +03:00
Evgeny Mankov 4a5484c616 [HIPIFY] All CUDA 8.0.44 API functions update
(for both Driver and Runtime APIs)

1) P2P
cuDeviceGetP2PAttribute   cudaDeviceGetP2PAttribute

2) Memory Mngmnt
cuMemPrefetchAsync        cudaMemPrefetchAsync
cuMemAdvise               cudaMemAdvise
cuMemRangeGetAttribute    cudaMemRangeGetAttribute
cuMemRangeGetAttributes   cudaMemRangeGetAttributes

3) Streams (Driver API only, no analogues in Runtime API)
cuStreamWaitValue32
cuStreamWaitValue32
cuStreamWriteValue32

4) Texture Reference Mngmnt (Driver API only, no analogues in Runtime API)
cuTexRefSetBorderColor
cuTexRefGetBorderColor
2017-06-01 21:08:33 +03:00
Evgeny Mankov 6e99e388ea [HIP] [HIPIFY] CUDA Driver API 8.0.44 JIT options support. 2017-05-31 18:55:29 +03:00
Evgeny Mankov 063539308e [HIPIFY] Add the rest CUDA Runtime API 8.0.44 Data structures.
+ sync with corresponding CUDA Driver API Data structures.

P.S.
There is no any new changes in CUDA Runtime API 8.0.61 Data structures since 8.0.44.
2017-05-30 19:45:59 +03:00
Evgeny Mankov 1cc5f42e34 [HIPIFY] Add the rest CUDA Driver API 8.0.44 Data structures.
+ Memory advise values
+ Memory Range Attributes
+ P2P Attributes

P.S.
There is no any new changes in CUDA Driver API 8.0.61 Data structures since 8.0.44.
2017-05-30 18:29:14 +03:00
Evgeny Mankov 3e99bc23e7 [HIPIFY] Add more CUDA Driver API 8.0.44 Data structures. 2017-05-30 17:58:13 +03:00
Evgeny Mankov c863215611 [FIX] [HIPIFY] Add matchers for function return types.
https://github.com/GPUOpen-ProfessionalCompute-Tools/HIP/issues/73

Examples (https://github.com/thrust/thrust/blob/master/thrust/system/cuda/detail/trivial_copy.inl):

template<typename System1,
         typename System2>
cudaStream_t cuda_memcpy_stream(const thrust::cpp::execution_policy<System1> &,
                                const thrust::cuda::execution_policy<System2> &exec)

template<typename System1,
         typename System2>
cudaMemcpyKind cuda_memcpy_kind(const thrust::cuda::execution_policy<System1> &,
                                const thrust::cpp::execution_policy<System2> &)
2017-05-24 18:25:40 +03:00
Evgeny Mankov 3d973dc4da [FIX] [HIPIFY] Matcher for new operator is missing.
https://github.com/GPUOpen-ProfessionalCompute-Tools/HIP/issues/80

Example from CUDA 8.0.44 sample (CUDASamples\0_Simple\matrixMulDrv\matrixMulDrv.cpp):
    CUjit_option *jitOptions = new CUjit_option[jitNumOptions];
where CUjit_option is enum, should be:
    hipJitOption *jitOptions = new hipJitOption[jitNumOptions];
2017-05-23 19:45:38 +03:00
Evgeny Mankov b5a1d47e68 [HIPIFY] [FIX] [HIPIFY] Matcher for pointer to enum var declaration is missing.
https://github.com/GPUOpen-ProfessionalCompute-Tools/HIP/issues/79

Example from CUDA 8.0.44 sample (CUDASamples\0_Simple\matrixMulDrv\matrixMulDrv.cpp):
    CUjit_option *jitOptions = new CUjit_option[jitNumOptions];
where CUjit_option is enum, should be:
    hipJitOption *jitOptions = new hipJitOption[jitNumOptions];

[TODO]
1. new CUjit_option -> new hipJitOption.
   Matcher for new operator is missing:
https://github.com/GPUOpen-ProfessionalCompute-Tools/HIP/issues/79
2. Merge matchers cudaEnumDecl and cudaEnumVarPtr.
2017-05-19 17:39:09 +03:00
Evgeny Mankov 270f643c9c [HIP] [HIPIFY] [FIX] cuModuleLoadDataEx -> hipModuleLoadDataEx
https://github.com/GPUOpen-ProfessionalCompute-Tools/HIP/issues/81

1.	Do not use JIT options on HCC path, call hipModuleLoadData instead.
2.	NVCC path is unchanged, to call cuModuleLoadDataEx with all options.
3.	Get rid of manual hipification, based on #ifdef #else for NVCC/HIP.
4.	Update documentation accordingly.
2017-05-19 17:22:14 +03:00
Evgeny Mankov 12d8c53c90 [HIPIFY] cudaMallocPitch -> hipMallocPitch 2017-05-16 18:21:25 +03:00
Evgeny Mankov a97cb6810c [HIPIFY] CUDA Driver API: Primary Context Management support. 2017-05-15 15:35:52 +03:00