Kokkos Core Kernels Package
Version of the Day
core
src
Kokkos_Macros.hpp
1
/*
2
//@HEADER
3
// ************************************************************************
4
//
5
// Kokkos v. 3.0
6
// Copyright (2020) National Technology & Engineering
7
// Solutions of Sandia, LLC (NTESS).
8
//
9
// Under the terms of Contract DE-NA0003525 with NTESS,
10
// the U.S. Government retains certain rights in this software.
11
//
12
// Redistribution and use in source and binary forms, with or without
13
// modification, are permitted provided that the following conditions are
14
// met:
15
//
16
// 1. Redistributions of source code must retain the above copyright
17
// notice, this list of conditions and the following disclaimer.
18
//
19
// 2. Redistributions in binary form must reproduce the above copyright
20
// notice, this list of conditions and the following disclaimer in the
21
// documentation and/or other materials provided with the distribution.
22
//
23
// 3. Neither the name of the Corporation nor the names of the
24
// contributors may be used to endorse or promote products derived from
25
// this software without specific prior written permission.
26
//
27
// THIS SOFTWARE IS PROVIDED BY NTESS "AS IS" AND ANY
28
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
29
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
30
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL NTESS OR THE
31
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
32
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
33
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
34
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
35
// LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
36
// NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
37
// SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
38
//
39
// Questions? Contact Christian R. Trott (crtrott@sandia.gov)
40
//
41
// ************************************************************************
42
//@HEADER
43
*/
44
45
#ifndef KOKKOS_MACROS_HPP
46
#define KOKKOS_MACROS_HPP
47
48
//----------------------------------------------------------------------------
63
#ifndef KOKKOS_DONT_INCLUDE_CORE_CONFIG_H
64
#include <KokkosCore_config.h>
65
#endif
66
67
//----------------------------------------------------------------------------
98
//----------------------------------------------------------------------------
99
100
#if !defined(KOKKOS_ENABLE_THREADS) && !defined(KOKKOS_ENABLE_CUDA) && \
101
!defined(KOKKOS_ENABLE_OPENMP) && !defined(KOKKOS_ENABLE_HPX) && \
102
!defined(KOKKOS_ENABLE_OPENMPTARGET) && !defined(KOKKOS_ENABLE_HIP) && \
103
!defined(KOKKOS_ENABLE_SYCL)
104
#define KOKKOS_INTERNAL_NOT_PARALLEL
105
#endif
106
107
#define KOKKOS_ENABLE_CXX11_DISPATCH_LAMBDA
108
109
#include <KokkosCore_Config_SetupBackend.hpp>
110
111
//----------------------------------------------------------------------------
112
// Mapping compiler built-ins to KOKKOS_COMPILER_*** macros
113
114
#if defined(__NVCC__)
115
// NVIDIA compiler is being used.
116
// Code is parsed and separated into host and device code.
117
// Host code is compiled again with another compiler.
118
// Device code is compile to 'ptx'.
119
#define KOKKOS_COMPILER_NVCC __NVCC__
120
#endif
// #if defined( __NVCC__ )
121
122
#if !defined(KOKKOS_LAMBDA)
123
#define KOKKOS_LAMBDA [=]
124
#endif
125
126
#if (defined(KOKKOS_ENABLE_CXX17) || defined(KOKKOS_ENABLE_CXX20)) && \
127
!defined(KOKKOS_CLASS_LAMBDA)
128
#define KOKKOS_CLASS_LAMBDA [ =, *this ]
129
#endif
130
131
//#if !defined( __CUDA_ARCH__ ) // Not compiling Cuda code to 'ptx'.
132
133
// Intel compiler for host code.
134
135
#if defined(__INTEL_COMPILER)
136
#define KOKKOS_COMPILER_INTEL __INTEL_COMPILER
137
#elif defined(__INTEL_LLVM_COMPILER)
138
#define KOKKOS_COMPILER_INTEL __INTEL_LLVM_COMPILER
139
#elif defined(__ICC)
140
// Old define
141
#define KOKKOS_COMPILER_INTEL __ICC
142
#elif defined(__ECC)
143
// Very old define
144
#define KOKKOS_COMPILER_INTEL __ECC
145
#endif
146
147
// CRAY compiler for host code
148
#if defined(_CRAYC)
149
#define KOKKOS_COMPILER_CRAYC _CRAYC
150
#endif
151
152
#if defined(__IBMCPP__)
153
// IBM C++
154
#define KOKKOS_COMPILER_IBM __IBMCPP__
155
#elif defined(__IBMC__)
156
#define KOKKOS_COMPILER_IBM __IBMC__
157
#elif defined(__ibmxl_vrm__)
// xlclang++
158
#define KOKKOS_COMPILER_IBM __ibmxl_vrm__
159
#endif
160
161
#if defined(__APPLE_CC__)
162
#define KOKKOS_COMPILER_APPLECC __APPLE_CC__
163
#endif
164
165
#if defined(__clang__) && !defined(KOKKOS_COMPILER_INTEL) && \
166
!defined(KOKKOS_COMPILER_IBM)
167
#define KOKKOS_COMPILER_CLANG \
168
__clang_major__ * 100 + __clang_minor__ * 10 + __clang_patchlevel__
169
#endif
170
171
#if !defined(__clang__) && !defined(KOKKOS_COMPILER_INTEL) && defined(__GNUC__)
172
#define KOKKOS_COMPILER_GNU \
173
__GNUC__ * 100 + __GNUC_MINOR__ * 10 + __GNUC_PATCHLEVEL__
174
175
#if (530 > KOKKOS_COMPILER_GNU)
176
#error "Compiling with GCC version earlier than 5.3.0 is not supported."
177
#endif
178
#endif
179
180
#if defined(__PGIC__)
181
#define KOKKOS_COMPILER_PGI \
182
__PGIC__ * 100 + __PGIC_MINOR__ * 10 + __PGIC_PATCHLEVEL__
183
184
#if (1740 > KOKKOS_COMPILER_PGI)
185
#error "Compiling with PGI version earlier than 17.4 is not supported."
186
#endif
187
#endif
188
189
#if defined(_MSC_VER) && !defined(KOKKOS_COMPILER_INTEL)
190
#define KOKKOS_COMPILER_MSVC _MSC_VER
191
#endif
192
193
#if defined(_OPENMP)
194
// Compiling with OpenMP.
195
// The value of _OPENMP is an integer value YYYYMM
196
// where YYYY and MM are the year and month designation
197
// of the supported OpenMP API version.
198
#endif
// #if defined( _OPENMP )
199
200
//----------------------------------------------------------------------------
201
// Intel compiler macros
202
203
#if defined(KOKKOS_COMPILER_INTEL)
204
// FIXME_SYCL
205
#if !defined(KOKKOS_ENABLE_SYCL)
206
#define KOKKOS_ENABLE_PRAGMA_UNROLL 1
207
#define KOKKOS_ENABLE_PRAGMA_LOOPCOUNT 1
208
#define KOKKOS_ENABLE_PRAGMA_VECTOR 1
209
#endif
210
#if (1800 > KOKKOS_COMPILER_INTEL)
211
#define KOKKOS_ENABLE_PRAGMA_SIMD 1
212
#endif
213
214
// FIXME_SYCL
215
#if !defined(KOKKOS_ENABLE_SYCL)
216
#define KOKKOS_ENABLE_PRAGMA_IVDEP 1
217
#endif
218
219
#if !defined(KOKKOS_MEMORY_ALIGNMENT)
220
#define KOKKOS_MEMORY_ALIGNMENT 64
221
#endif
222
223
#define KOKKOS_RESTRICT __restrict__
224
225
#ifndef KOKKOS_IMPL_ALIGN_PTR
226
#define KOKKOS_IMPL_ALIGN_PTR(size) __attribute__((align_value(size)))
227
#endif
228
229
#if (1700 > KOKKOS_COMPILER_INTEL)
230
#error "Compiling with Intel version earlier than 17.0 is not supported."
231
#endif
232
233
#if !defined(KOKKOS_ENABLE_ASM) && !defined(_WIN32)
234
#define KOKKOS_ENABLE_ASM 1
235
#endif
236
237
#if !defined(KOKKOS_IMPL_FORCEINLINE_FUNCTION)
238
#if !defined(_WIN32)
239
#define KOKKOS_IMPL_FORCEINLINE_FUNCTION inline __attribute__((always_inline))
240
#define KOKKOS_IMPL_FORCEINLINE __attribute__((always_inline))
241
#else
242
#define KOKKOS_IMPL_FORCEINLINE_FUNCTION inline
243
#endif
244
#endif
245
246
#if defined(KOKKOS_ARCH_AVX512MIC)
247
#define KOKKOS_ENABLE_RFO_PREFETCH 1
248
#if (KOKKOS_COMPILER_INTEL < 1800) && !defined(KOKKOS_KNL_USE_ASM_WORKAROUND)
249
#define KOKKOS_KNL_USE_ASM_WORKAROUND 1
250
#endif
251
#endif
252
253
#if (1800 > KOKKOS_COMPILER_INTEL)
254
#define KOKKOS_IMPL_INTEL_WORKAROUND_NOEXCEPT_SPECIFICATION_VIRTUAL_FUNCTION
255
#endif
256
257
#if defined(__MIC__)
258
// Compiling for Xeon Phi
259
#endif
260
#endif
261
262
//----------------------------------------------------------------------------
263
// Cray compiler macros
264
265
#if defined(KOKKOS_COMPILER_CRAYC)
266
#endif
267
268
//----------------------------------------------------------------------------
269
// IBM Compiler macros
270
271
#if defined(KOKKOS_COMPILER_IBM)
272
#define KOKKOS_ENABLE_PRAGMA_UNROLL 1
273
//#define KOKKOS_ENABLE_PRAGMA_IVDEP 1
274
//#define KOKKOS_ENABLE_PRAGMA_LOOPCOUNT 1
275
//#define KOKKOS_ENABLE_PRAGMA_VECTOR 1
276
//#define KOKKOS_ENABLE_PRAGMA_SIMD 1
277
278
#if !defined(KOKKOS_ENABLE_ASM)
279
#define KOKKOS_ENABLE_ASM 1
280
#endif
281
#endif
282
283
//----------------------------------------------------------------------------
284
// CLANG compiler macros
285
286
#if defined(KOKKOS_COMPILER_CLANG)
287
//#define KOKKOS_ENABLE_PRAGMA_UNROLL 1
288
//#define KOKKOS_ENABLE_PRAGMA_IVDEP 1
289
//#define KOKKOS_ENABLE_PRAGMA_LOOPCOUNT 1
290
//#define KOKKOS_ENABLE_PRAGMA_VECTOR 1
291
//#define KOKKOS_ENABLE_PRAGMA_SIMD 1
292
293
#if !defined(KOKKOS_IMPL_FORCEINLINE_FUNCTION)
294
#define KOKKOS_IMPL_FORCEINLINE_FUNCTION inline __attribute__((always_inline))
295
#define KOKKOS_IMPL_FORCEINLINE __attribute__((always_inline))
296
#endif
297
298
#if !defined(KOKKOS_IMPL_ALIGN_PTR)
299
#define KOKKOS_IMPL_ALIGN_PTR(size) __attribute__((aligned(size)))
300
#endif
301
302
#endif
303
304
//----------------------------------------------------------------------------
305
// GNU Compiler macros
306
307
#if defined(KOKKOS_COMPILER_GNU)
308
//#define KOKKOS_ENABLE_PRAGMA_UNROLL 1
309
//#define KOKKOS_ENABLE_PRAGMA_IVDEP 1
310
//#define KOKKOS_ENABLE_PRAGMA_LOOPCOUNT 1
311
//#define KOKKOS_ENABLE_PRAGMA_VECTOR 1
312
//#define KOKKOS_ENABLE_PRAGMA_SIMD 1
313
314
#if defined(KOKKOS_ARCH_AVX512MIC)
315
#define KOKKOS_ENABLE_RFO_PREFETCH 1
316
#endif
317
318
#if !defined(KOKKOS_IMPL_FORCEINLINE_FUNCTION)
319
#define KOKKOS_IMPL_FORCEINLINE_FUNCTION inline __attribute__((always_inline))
320
#define KOKKOS_IMPL_FORCEINLINE __attribute__((always_inline))
321
#endif
322
323
#define KOKKOS_RESTRICT __restrict__
324
325
#if !defined(KOKKOS_ENABLE_ASM) && !defined(__PGIC__) && \
326
(defined(__amd64) || defined(__amd64__) || defined(__x86_64) || \
327
defined(__x86_64__) || defined(__PPC64__))
328
#define KOKKOS_ENABLE_ASM 1
329
#endif
330
#endif
331
332
//----------------------------------------------------------------------------
333
334
#if defined(KOKKOS_COMPILER_PGI)
335
#define KOKKOS_ENABLE_PRAGMA_UNROLL 1
336
#define KOKKOS_ENABLE_PRAGMA_IVDEP 1
337
//#define KOKKOS_ENABLE_PRAGMA_LOOPCOUNT 1
338
#define KOKKOS_ENABLE_PRAGMA_VECTOR 1
339
//#define KOKKOS_ENABLE_PRAGMA_SIMD 1
340
#endif
341
342
//----------------------------------------------------------------------------
343
344
#if defined(KOKKOS_COMPILER_NVCC)
345
#if defined(__CUDA_ARCH__)
346
#define KOKKOS_ENABLE_PRAGMA_UNROLL 1
347
#endif
348
#endif
349
350
//----------------------------------------------------------------------------
351
// Define function marking macros if compiler specific macros are undefined:
352
353
#if !defined(KOKKOS_IMPL_FORCEINLINE_FUNCTION)
354
#define KOKKOS_IMPL_FORCEINLINE_FUNCTION inline
355
#endif
356
357
#if !defined(KOKKOS_IMPL_FORCEINLINE)
358
#define KOKKOS_IMPL_FORCEINLINE inline
359
#endif
360
361
#if !defined(KOKKOS_IMPL_INLINE_FUNCTION)
362
#define KOKKOS_IMPL_INLINE_FUNCTION inline
363
#endif
364
365
#if !defined(KOKKOS_IMPL_FUNCTION)
366
#define KOKKOS_IMPL_FUNCTION
367
#endif
368
369
#if !defined(KOKKOS_INLINE_FUNCTION_DELETED)
370
#define KOKKOS_INLINE_FUNCTION_DELETED inline
371
#endif
372
373
#if !defined(KOKKOS_DEFAULTED_FUNCTION)
374
#define KOKKOS_DEFAULTED_FUNCTION inline
375
#endif
376
377
#if !defined(KOKKOS_IMPL_HOST_FUNCTION)
378
#define KOKKOS_IMPL_HOST_FUNCTION
379
#endif
380
381
#if !defined(KOKKOS_IMPL_DEVICE_FUNCTION)
382
#define KOKKOS_IMPL_DEVICE_FUNCTION
383
#endif
384
385
// Temporary solution for SYCL not supporting printf in kernels.
386
// Might disappear at any point once we have found another solution.
387
#if !defined(KOKKOS_IMPL_DO_NOT_USE_PRINTF)
388
#define KOKKOS_IMPL_DO_NOT_USE_PRINTF(...) printf(__VA_ARGS__)
389
#endif
390
391
//----------------------------------------------------------------------------
392
// Define final version of functions. This is so that clang tidy can find these
393
// macros more easily
394
#if defined(__clang_analyzer__)
395
#define KOKKOS_FUNCTION \
396
KOKKOS_IMPL_FUNCTION __attribute__((annotate("KOKKOS_FUNCTION"
)))
397
#define KOKKOS_INLINE_FUNCTION \
398
KOKKOS_IMPL_INLINE_FUNCTION \
399
__attribute__((annotate("KOKKOS_INLINE_FUNCTION"
)))
400
#define KOKKOS_FORCEINLINE_FUNCTION \
401
KOKKOS_IMPL_FORCEINLINE_FUNCTION \
402
__attribute__((annotate("KOKKOS_FORCEINLINE_FUNCTION"
)))
403
#else
404
#define KOKKOS_FUNCTION KOKKOS_IMPL_FUNCTION
405
#define KOKKOS_INLINE_FUNCTION KOKKOS_IMPL_INLINE_FUNCTION
406
#define KOKKOS_FORCEINLINE_FUNCTION KOKKOS_IMPL_FORCEINLINE_FUNCTION
407
#endif
408
409
//----------------------------------------------------------------------------
410
// Define empty macro for restrict if necessary:
411
412
#if !defined(KOKKOS_RESTRICT)
413
#define KOKKOS_RESTRICT
414
#endif
415
416
//----------------------------------------------------------------------------
417
// Define Macro for alignment:
418
419
#if !defined(KOKKOS_MEMORY_ALIGNMENT)
420
#define KOKKOS_MEMORY_ALIGNMENT 64
421
#endif
422
423
#if !defined(KOKKOS_MEMORY_ALIGNMENT_THRESHOLD)
424
#define KOKKOS_MEMORY_ALIGNMENT_THRESHOLD 1
425
#endif
426
427
#if !defined(KOKKOS_IMPL_ALIGN_PTR)
428
#define KOKKOS_IMPL_ALIGN_PTR(size)
/* */
429
#endif
430
431
//----------------------------------------------------------------------------
432
// Determine the default execution space for parallel dispatch.
433
// There is zero or one default execution space specified.
434
435
#if 1 < ((defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_CUDA) ? 1 : 0) + \
436
(defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_HIP) ? 1 : 0) + \
437
(defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_SYCL) ? 1 : 0) + \
438
(defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_OPENMPTARGET) ? 1 : 0) + \
439
(defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_OPENMP) ? 1 : 0) + \
440
(defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_THREADS) ? 1 : 0) + \
441
(defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_HPX) ? 1 : 0) + \
442
(defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_SERIAL) ? 1 : 0))
443
#error "More than one KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_* specified."
444
#endif
445
446
// If default is not specified then chose from enabled execution spaces.
447
// Priority: CUDA, HIP, SYCL, OPENMPTARGET, OPENMP, THREADS, HPX, SERIAL
448
#if defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_CUDA)
449
#elif defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_HIP)
450
#elif defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_SYCL)
451
#elif defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_OPENMPTARGET)
452
#elif defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_OPENMP)
453
#elif defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_THREADS)
454
#elif defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_HPX)
455
#elif defined(KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_SERIAL)
456
#elif defined(KOKKOS_ENABLE_CUDA)
457
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_CUDA
458
#elif defined(KOKKOS_ENABLE_HIP)
459
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_HIP
460
#if defined(__HIP__)
461
// mark that HIP-clang can use __host__ and __device__
462
// as valid overload criteria
463
#define KOKKOS_IMPL_ENABLE_OVERLOAD_HOST_DEVICE
464
#endif
465
#elif defined(KOKKOS_ENABLE_SYCL)
466
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_SYCL
467
#elif defined(KOKKOS_ENABLE_OPENMPTARGET)
468
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_OPENMPTARGET
469
#elif defined(KOKKOS_ENABLE_OPENMP)
470
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_OPENMP
471
#elif defined(KOKKOS_ENABLE_THREADS)
472
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_THREADS
473
#elif defined(KOKKOS_ENABLE_HPX)
474
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_HPX
475
#else
476
#define KOKKOS_ENABLE_DEFAULT_DEVICE_TYPE_SERIAL
477
#endif
478
479
//----------------------------------------------------------------------------
480
// Determine for what space the code is being compiled:
481
482
#if defined(__CUDACC__) && defined(__CUDA_ARCH__) && defined(KOKKOS_ENABLE_CUDA)
483
#define KOKKOS_ACTIVE_EXECUTION_MEMORY_SPACE_CUDA
484
#elif defined(__SYCL_DEVICE_ONLY__) && defined(KOKKOS_ENABLE_SYCL)
485
#define KOKKOS_ACTIVE_EXECUTION_MEMORY_SPACE_SYCL
486
#elif defined(__HIPCC__) && defined(__HIP_DEVICE_COMPILE__) && \
487
defined(KOKKOS_ENABLE_HIP)
488
#define KOKKOS_ACTIVE_EXECUTION_MEMORY_SPACE_HIP_GPU
489
#else
490
#define KOKKOS_ACTIVE_EXECUTION_MEMORY_SPACE_HOST
491
#endif
492
493
//----------------------------------------------------------------------------
494
495
#if (defined(_POSIX_C_SOURCE) && _POSIX_C_SOURCE >= 200112L) || \
496
(defined(_XOPEN_SOURCE) && _XOPEN_SOURCE >= 600)
497
#if defined(KOKKOS_ENABLE_PERFORMANCE_POSIX_MEMALIGN)
498
#define KOKKOS_ENABLE_POSIX_MEMALIGN 1
499
#endif
500
#endif
501
502
//----------------------------------------------------------------------------
503
// If compiling with CUDA, we must use relocateable device code
504
// to enable the task policy.
505
506
#if defined(KOKKOS_ENABLE_CUDA)
507
#if defined(KOKKOS_ENABLE_CUDA_RELOCATABLE_DEVICE_CODE)
508
#define KOKKOS_ENABLE_TASKDAG
509
#endif
510
#elif !defined(KOKKOS_ENABLE_HIP) && !defined(KOKKOS_ENABLE_SYCL)
511
#define KOKKOS_ENABLE_TASKDAG
512
#endif
513
514
#if defined(KOKKOS_ENABLE_CUDA)
515
#define KOKKOS_IMPL_CUDA_VERSION_9_WORKAROUND
516
#if (__CUDA_ARCH__)
517
#define KOKKOS_IMPL_CUDA_SYNCWARP_NEEDS_MASK
518
#endif
519
#endif
520
521
#define KOKKOS_INVALID_INDEX (~std::size_t(0))
522
523
#define KOKKOS_IMPL_CTOR_DEFAULT_ARG KOKKOS_INVALID_INDEX
524
525
#define KOKKOS_CONSTEXPR_14 constexpr
526
#define KOKKOS_DEPRECATED [[deprecated]]
527
#define KOKKOS_DEPRECATED_TRAILING_ATTRIBUTE
528
529
// DJS 05/28/2019: Bugfix: Issue 2155
530
// Use KOKKOS_ENABLE_CUDA_LDG_INTRINSIC to avoid memory leak in RandomAccess
531
// View
532
#if defined(KOKKOS_ENABLE_CUDA) && !defined(KOKKOS_ENABLE_CUDA_LDG_INTRINSIC)
533
#define KOKKOS_ENABLE_CUDA_LDG_INTRINSIC
534
#endif
535
536
#if defined(KOKKOS_ENABLE_CXX17) || defined(KOKKOS_ENABLE_CXX20)
537
#define KOKKOS_ATTRIBUTE_NODISCARD [[nodiscard]]
538
#else
539
#define KOKKOS_ATTRIBUTE_NODISCARD
540
#endif
541
542
#if (defined(KOKKOS_COMPILER_GNU) || defined(KOKKOS_COMPILER_CLANG) || \
543
defined(KOKKOS_COMPILER_INTEL) || defined(KOKKOS_COMPILER_PGI)) && \
544
!defined(KOKKOS_COMPILER_MSVC)
545
#define KOKKOS_IMPL_ENABLE_STACKTRACE
546
#define KOKKOS_IMPL_ENABLE_CXXABI
547
#endif
548
549
// WORKAROUND for AMD aomp which apparently defines CUDA_ARCH when building for
550
// AMD GPUs with OpenMP Target ???
551
#if defined(__CUDA_ARCH__) && !defined(__CUDACC__) && \
552
!defined(KOKKOS_ENABLE_HIP) && !defined(KOKKOS_ENABLE_CUDA)
553
#undef __CUDA_ARCH__
554
#endif
555
556
#if defined(KOKKOS_COMPILER_MSVC) && !defined(KOKKOS_COMPILER_CLANG)
557
#define KOKKOS_THREAD_LOCAL __declspec(thread)
558
#else
559
#define KOKKOS_THREAD_LOCAL __thread
560
#endif
561
562
#if (defined(KOKKOS_IMPL_WINDOWS_CUDA) || defined(KOKKOS_COMPILER_MSVC)) && \
563
!defined(KOKKOS_COMPILER_CLANG)
564
// MSVC (as of 16.5.5 at least) does not do empty base class optimization by
565
// default when there are multiple bases, even though the standard requires it
566
// for standard layout types.
567
#define KOKKOS_IMPL_ENFORCE_EMPTY_BASE_OPTIMIZATION __declspec(empty_bases)
568
#else
569
#define KOKKOS_IMPL_ENFORCE_EMPTY_BASE_OPTIMIZATION
570
#endif
571
572
#endif
// #ifndef KOKKOS_MACROS_HPP
Generated by
1.9.4
1.9.4