Coverage for cuda/bindings/nvrtc.pyx: 49.13%

1# SPDX-FileCopyrightText: Copyright (c) 2021-2026 NVIDIA CORPORATION & AFFILIATES. All rights reserved. (empty)

2# SPDX-License-Identifier: Apache-2.0 

3  

4# This code was automatically generated with version 13.3.0. Do not modify it directly. 

5# CYTHON-BINDINGS-GENERATED-DO-NOT-MODIFY-THIS-FILE: format=1; content-sha256=9c30b7a3d79561c6721e7df7b758eaee6afdd2763d9692ff638815abb64417e0 

6from typing import Any, Optional (empty)

7import cython 

8import ctypes (empty)

9from libc.stdlib cimport calloc, malloc, free 

10from libc cimport string 

11from libc.stdint cimport int32_t, uint32_t, int64_t, uint64_t, uintptr_t 

12from libc.stddef cimport wchar_t 

13from libc.limits cimport CHAR_MIN 

14from libcpp.vector cimport vector 

15from cpython.buffer cimport PyObject_CheckBuffer, PyObject_GetBuffer, PyBuffer_Release, PyBUF_SIMPLE, PyBUF_ANY_CONTIGUOUS 

16from cpython.bytes cimport PyBytes_FromStringAndSize 

17from ._internal._fast_enum import FastEnum as _FastEnum (empty)

18  

19import cuda.bindings.driver as _driver (empty)

20_driver = _driver.__dict__ (empty)

21include "_lib/utils.pxi" 

22  

23ctypedef unsigned long long signed_char_ptr 

24ctypedef unsigned long long unsigned_char_ptr 

25ctypedef unsigned long long char_ptr 

26ctypedef unsigned long long short_ptr 

27ctypedef unsigned long long unsigned_short_ptr 

28ctypedef unsigned long long int_ptr 

29ctypedef unsigned long long long_int_ptr 

30ctypedef unsigned long long long_long_int_ptr 

31ctypedef unsigned long long unsigned_int_ptr 

32ctypedef unsigned long long unsigned_long_int_ptr 

33ctypedef unsigned long long unsigned_long_long_int_ptr 

34ctypedef unsigned long long uint32_t_ptr 

35ctypedef unsigned long long uint64_t_ptr 

36ctypedef unsigned long long int32_t_ptr 

37ctypedef unsigned long long int64_t_ptr 

38ctypedef unsigned long long unsigned_ptr 

39ctypedef unsigned long long unsigned_long_long_ptr 

40ctypedef unsigned long long long_long_ptr 

41ctypedef unsigned long long size_t_ptr 

42ctypedef unsigned long long long_ptr 

43ctypedef unsigned long long float_ptr 

44ctypedef unsigned long long double_ptr 

45ctypedef unsigned long long void_ptr 

46  

47#: Flags for nvrtcInstallBundledHeaders. 

48#: 

49#: Skip installation if version marker exists and version matches. This is 

50#: the default behavior when flags=0. 

51NVRTC_INSTALL_HEADERS_SKIP_IF_EXISTS = cynvrtc.NVRTC_INSTALL_HEADERS_SKIP_IF_EXISTS (empty)

52  

53#: Clear existing directory contents before installation. Guarantees 

54#: consistency by removing any existing files first. 

55NVRTC_INSTALL_HEADERS_FORCE_OVERWRITE = cynvrtc.NVRTC_INSTALL_HEADERS_FORCE_OVERWRITE (empty)

56  

57#: Return NVRTC_ERROR_BUSY immediately if installation is in progress by 

58#: another process, instead of waiting for the lock. Can be combined with 

59#: FORCE_OVERWRITE using bitwise OR. 

60NVRTC_INSTALL_HEADERS_NO_WAIT = cynvrtc.NVRTC_INSTALL_HEADERS_NO_WAIT (empty)

61  

62class nvrtcResult(_FastEnum): (empty)

63 """ 

64 The enumerated type :py:obj:`~.nvrtcResult` defines API call result 

65 codes. NVRTC API functions return :py:obj:`~.nvrtcResult` to 

66 indicate the call result. 

67 """ 

68  

69 NVRTC_SUCCESS = cynvrtc.nvrtcResult.NVRTC_SUCCESS (empty)

70  

71 NVRTC_ERROR_OUT_OF_MEMORY = cynvrtc.nvrtcResult.NVRTC_ERROR_OUT_OF_MEMORY (empty)

72  

73 NVRTC_ERROR_PROGRAM_CREATION_FAILURE = cynvrtc.nvrtcResult.NVRTC_ERROR_PROGRAM_CREATION_FAILURE (empty)

74  

75 NVRTC_ERROR_INVALID_INPUT = cynvrtc.nvrtcResult.NVRTC_ERROR_INVALID_INPUT (empty)

76  

77 NVRTC_ERROR_INVALID_PROGRAM = cynvrtc.nvrtcResult.NVRTC_ERROR_INVALID_PROGRAM (empty)

78  

79 NVRTC_ERROR_INVALID_OPTION = cynvrtc.nvrtcResult.NVRTC_ERROR_INVALID_OPTION (empty)

80  

81 NVRTC_ERROR_COMPILATION = cynvrtc.nvrtcResult.NVRTC_ERROR_COMPILATION (empty)

82  

83 NVRTC_ERROR_BUILTIN_OPERATION_FAILURE = cynvrtc.nvrtcResult.NVRTC_ERROR_BUILTIN_OPERATION_FAILURE (empty)

84  

85 NVRTC_ERROR_NO_NAME_EXPRESSIONS_AFTER_COMPILATION = cynvrtc.nvrtcResult.NVRTC_ERROR_NO_NAME_EXPRESSIONS_AFTER_COMPILATION (empty)

86  

87 NVRTC_ERROR_NO_LOWERED_NAMES_BEFORE_COMPILATION = cynvrtc.nvrtcResult.NVRTC_ERROR_NO_LOWERED_NAMES_BEFORE_COMPILATION (empty)

88  

89 NVRTC_ERROR_NAME_EXPRESSION_NOT_VALID = cynvrtc.nvrtcResult.NVRTC_ERROR_NAME_EXPRESSION_NOT_VALID (empty)

90  

91 NVRTC_ERROR_INTERNAL_ERROR = cynvrtc.nvrtcResult.NVRTC_ERROR_INTERNAL_ERROR (empty)

92  

93 NVRTC_ERROR_TIME_FILE_WRITE_FAILED = cynvrtc.nvrtcResult.NVRTC_ERROR_TIME_FILE_WRITE_FAILED (empty)

94  

95 NVRTC_ERROR_NO_PCH_CREATE_ATTEMPTED = cynvrtc.nvrtcResult.NVRTC_ERROR_NO_PCH_CREATE_ATTEMPTED (empty)

96  

97 NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED = cynvrtc.nvrtcResult.NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED (empty)

98  

99 NVRTC_ERROR_PCH_CREATE = cynvrtc.nvrtcResult.NVRTC_ERROR_PCH_CREATE (empty)

100  

101 NVRTC_ERROR_CANCELLED = cynvrtc.nvrtcResult.NVRTC_ERROR_CANCELLED (empty)

102  

103 NVRTC_ERROR_TIME_TRACE_FILE_WRITE_FAILED = cynvrtc.nvrtcResult.NVRTC_ERROR_TIME_TRACE_FILE_WRITE_FAILED (empty)

104  

105 NVRTC_ERROR_BUSY = cynvrtc.nvrtcResult.NVRTC_ERROR_BUSY (empty)

106  

107cdef object _nvrtcResult = nvrtcResult (empty)

108cdef object _nvrtcResult_SUCCESS = nvrtcResult.NVRTC_SUCCESS (empty)

109  

110cdef class nvrtcProgram: 

111 """ nvrtcProgram is the unit of compilation, and an opaque handle for a program. 

112  

113 To compile a CUDA program string, an instance of nvrtcProgram must be created first with nvrtcCreateProgram, then compiled with nvrtcCompileProgram. 

114  

115 Methods 

116 ------- 

117 getPtr() 

118 Get memory address of class instance 

119  

120 """ 

121 def __cinit__(self, void_ptr init_value = 0, void_ptr _ptr = 0): 

122 if _ptr == 0: 31 ctx2b c d e f g h i k l m n o p q r s t u v w x y z A B D j }b~bac

123 self._pvt_ptr = &self._pvt_val 31 ctx2b c d e f g h i k l m n o p q r s t u v w x y z A B D j }b~bac

124 self._pvt_ptr[0] = <cynvrtc.nvrtcProgram>init_value 31 ctx2b c d e f g h i k l m n o p q r s t u v w x y z A B D j }b~bac

125 else: 

126 self._pvt_ptr = <cynvrtc.nvrtcProgram *>_ptr 

127 def __init__(self, *args, **kwargs): 

128 pass 31 ctx2b c d e f g h i k l m n o p q r s t u v w x y z A B D j }b~bac

129 def __repr__(self): 

130 return '<nvrtcProgram ' + str(hex(self.__int__())) + '>' 

131 def __index__(self): 

132 return self.__int__() 

133 def __eq__(self, other): 

134 if not isinstance(other, nvrtcProgram): 

135 return False 

136 return self._pvt_ptr[0] == (<nvrtcProgram>other)._pvt_ptr[0] 

137 def __hash__(self): 

138 return hash(<uintptr_t><void*>(self._pvt_ptr[0])) 

139 def __int__(self): 

140 return <void_ptr>self._pvt_ptr[0] 28 ctx1bcdefghiklmnopqrstuvwxyzABDj

141 def getPtr(self): (empty)

142 return <void_ptr>self._pvt_ptr 19 ctx1klmnopqrstuvwxyzABj

143  

144cdef class nvrtcBundledHeadersInfo: 

145 """ 

146 Structure containing information about bundled headers. 

147  

148 Attributes 

149 ---------- 

150  

151 available : int 

152 Non-zero if bundled headers are available 

153  

154  

155 compressedSize : size_t 

156 Size of compressed archive in bytes 

157  

158  

159 uncompressedSize : size_t 

160 Estimated size when extracted in bytes 

161  

162  

163 cudaVersionMajor : int 

164 CUDA major version of bundled headers 

165  

166  

167 cudaVersionMinor : int 

168 CUDA minor version of bundled headers 

169  

170  

171 numFiles : unsigned int 

172 Number of header files in the bundle 

173  

174  

175 Methods 

176 ------- 

177 getPtr() 

178 Get memory address of class instance 

179 """ 

180 def __cinit__(self, void_ptr _ptr = 0): 

181 if _ptr == 0: 1 ctx1C

182 self._pvt_ptr = &self._pvt_val 1 ctx1C

183 else: 

184 self._pvt_ptr = <cynvrtc.nvrtcBundledHeadersInfo *>_ptr 

185 def __init__(self, void_ptr _ptr = 0): 

186 pass 1 ctx1C

187 def __dealloc__(self): 

188 pass 1 ctx1C

189 def getPtr(self): (empty)

190 return <void_ptr>self._pvt_ptr 

191 def __repr__(self): 

192 if self._pvt_ptr is not NULL: (empty)

193 str_list = [] 

194  

195 try: 

196 str_list += ['available : ' + str(self.available)] 

197 except ValueError: 

198 str_list += ['available : <ValueError>'] 

199  

200  

201 try: 

202 str_list += ['compressedSize : ' + str(self.compressedSize)] 

203 except ValueError: 

204 str_list += ['compressedSize : <ValueError>'] 

205  

206  

207 try: 

208 str_list += ['uncompressedSize : ' + str(self.uncompressedSize)] 

209 except ValueError: 

210 str_list += ['uncompressedSize : <ValueError>'] 

211  

212  

213 try: 

214 str_list += ['cudaVersionMajor : ' + str(self.cudaVersionMajor)] 

215 except ValueError: 

216 str_list += ['cudaVersionMajor : <ValueError>'] 

217  

218  

219 try: 

220 str_list += ['cudaVersionMinor : ' + str(self.cudaVersionMinor)] 

221 except ValueError: 

222 str_list += ['cudaVersionMinor : <ValueError>'] 

223  

224  

225 try: 

226 str_list += ['numFiles : ' + str(self.numFiles)] 

227 except ValueError: 

228 str_list += ['numFiles : <ValueError>'] 

229  

230 return '\n'.join(str_list) 

231 else: 

232 return '' 

233  

234 @property 

235 def available(self): 

236 return self._pvt_ptr[0].available 1 ctx1C

237 @available.setter 

238 def available(self, int available): 

239 self._pvt_ptr[0].available = available 

240  

241  

242 @property 

243 def compressedSize(self): 

244 return self._pvt_ptr[0].compressedSize 1 ctx1C

245 @compressedSize.setter 

246 def compressedSize(self, size_t compressedSize): 

247 self._pvt_ptr[0].compressedSize = compressedSize 

248  

249  

250 @property 

251 def uncompressedSize(self): 

252 return self._pvt_ptr[0].uncompressedSize 1 ctx1C

253 @uncompressedSize.setter 

254 def uncompressedSize(self, size_t uncompressedSize): 

255 self._pvt_ptr[0].uncompressedSize = uncompressedSize 

256  

257  

258 @property 

259 def cudaVersionMajor(self): 

260 return self._pvt_ptr[0].cudaVersionMajor 1 ctx1C

261 @cudaVersionMajor.setter 

262 def cudaVersionMajor(self, int cudaVersionMajor): 

263 self._pvt_ptr[0].cudaVersionMajor = cudaVersionMajor 

264  

265  

266 @property 

267 def cudaVersionMinor(self): 

268 return self._pvt_ptr[0].cudaVersionMinor 1 ctx1C

269 @cudaVersionMinor.setter 

270 def cudaVersionMinor(self, int cudaVersionMinor): 

271 self._pvt_ptr[0].cudaVersionMinor = cudaVersionMinor 

272  

273  

274 @property 

275 def numFiles(self): 

276 return self._pvt_ptr[0].numFiles 1 ctx1C

277 @numFiles.setter 

278 def numFiles(self, unsigned int numFiles): 

279 self._pvt_ptr[0].numFiles = numFiles 

280  

281  

282@cython.embedsignature(True) (empty)

283def nvrtcGetErrorString(result not None : nvrtcResult): 

284 """ nvrtcGetErrorString is a helper function that returns a string describing the given :py:obj:`~.nvrtcResult` code, e.g., NVRTC_SUCCESS to `"NVRTC_SUCCESS"`. For unrecognized enumeration values, it returns `"NVRTC_ERROR unknown"`. 

285  

286 Parameters 

287 ---------- 

288 result : :py:obj:`~.nvrtcResult` 

289 CUDA Runtime Compilation API result code. 

290  

291 Returns 

292 ------- 

293 nvrtcResult.NVRTC_SUCCESS 

294 nvrtcResult.NVRTC_SUCCESS 

295 bytes 

296 Message string for the given :py:obj:`~.nvrtcResult` code. 

297 """ 

298 cdef cynvrtc.nvrtcResult cyresult = int(result) 2 ctx2j bc

299 with nogil: 2 ctx2j bc

300 err = cynvrtc.nvrtcGetErrorString(cyresult) 2 ctx2j bc

301 return (nvrtcResult.NVRTC_SUCCESS, err) 2 ctx2j bc

302  

303@cython.embedsignature(True) (empty)

304def nvrtcVersion(): 

305 """ nvrtcVersion sets the output parameters `major` and `minor` with the CUDA Runtime Compilation version number. 

306  

307 Returns 

308 ------- 

309 nvrtcResult 

310 - :py:obj:`~.NVRTC_SUCCESS` 

311 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

312 major : int 

313 CUDA Runtime Compilation major version number. 

314 minor : int 

315 CUDA Runtime Compilation minor version number. 

316 """ 

317 cdef int major = 0 154 ctx2a b c d e f g h i F G H I J K L M N O P Q R S T U V W X Y Z 0 1 2 3 4 5 6 7 8 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | } ~ abbbcbdbebfbgbhbibjbkblbmbnbobpbqbrbsbtbubvbwbxbybzbAbBbCbDbEbFbGbHbIbJbKbLbMbNbObPbQbRbSbTbUbVbWbXbYbZb0b1b2b3b4b5b6b7b8b9b!b#b$b%b'b(b)b*b+b,b-b.b/b:b;b=b?b@b[b]b^b_b`b{b|b

318 cdef int minor = 0 154 ctx2a b c d e f g h i F G H I J K L M N O P Q R S T U V W X Y Z 0 1 2 3 4 5 6 7 8 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | } ~ abbbcbdbebfbgbhbibjbkblbmbnbobpbqbrbsbtbubvbwbxbybzbAbBbCbDbEbFbGbHbIbJbKbLbMbNbObPbQbRbSbTbUbVbWbXbYbZb0b1b2b3b4b5b6b7b8b9b!b#b$b%b'b(b)b*b+b,b-b.b/b:b;b=b?b@b[b]b^b_b`b{b|b

319 with nogil: 154 ctx2a b c d e f g h i F G H I J K L M N O P Q R S T U V W X Y Z 0 1 2 3 4 5 6 7 8 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | } ~ abbbcbdbebfbgbhbibjbkblbmbnbobpbqbrbsbtbubvbwbxbybzbAbBbCbDbEbFbGbHbIbJbKbLbMbNbObPbQbRbSbTbUbVbWbXbYbZb0b1b2b3b4b5b6b7b8b9b!b#b$b%b'b(b)b*b+b,b-b.b/b:b;b=b?b@b[b]b^b_b`b{b|b

320 err = cynvrtc.nvrtcVersion(&major, &minor) 154 ctx2a b c d e f g h i F G H I J K L M N O P Q R S T U V W X Y Z 0 1 2 3 4 5 6 7 8 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | } ~ abbbcbdbebfbgbhbibjbkblbmbnbobpbqbrbsbtbubvbwbxbybzbAbBbCbDbEbFbGbHbIbJbKbLbMbNbObPbQbRbSbTbUbVbWbXbYbZb0b1b2b3b4b5b6b7b8b9b!b#b$b%b'b(b)b*b+b,b-b.b/b:b;b=b?b@b[b]b^b_b`b{b|b

321 if err != cynvrtc.NVRTC_SUCCESS: 154 ctx2a b c d e f g h i F G H I J K L M N O P Q R S T U V W X Y Z 0 1 2 3 4 5 6 7 8 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | } ~ abbbcbdbebfbgbhbibjbkblbmbnbobpbqbrbsbtbubvbwbxbybzbAbBbCbDbEbFbGbHbIbJbKbLbMbNbObPbQbRbSbTbUbVbWbXbYbZb0b1b2b3b4b5b6b7b8b9b!b#b$b%b'b(b)b*b+b,b-b.b/b:b;b=b?b@b[b]b^b_b`b{b|b

322 return (_nvrtcResult(err), None, None) 

323 return (_nvrtcResult_SUCCESS, major, minor) 154 ctx2a b c d e f g h i F G H I J K L M N O P Q R S T U V W X Y Z 0 1 2 3 4 5 6 7 8 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | } ~ abbbcbdbebfbgbhbibjbkblbmbnbobpbqbrbsbtbubvbwbxbybzbAbBbCbDbEbFbGbHbIbJbKbLbMbNbObPbQbRbSbTbUbVbWbXbYbZb0b1b2b3b4b5b6b7b8b9b!b#b$b%b'b(b)b*b+b,b-b.b/b:b;b=b?b@b[b]b^b_b`b{b|b

324  

325@cython.embedsignature(True) (empty)

326def nvrtcGetNumSupportedArchs(): 

327 """ nvrtcGetNumSupportedArchs sets the output parameter `numArchs` with the number of architectures supported by NVRTC. This can then be used to pass an array to :py:obj:`~.nvrtcGetSupportedArchs` to get the supported architectures. 

328  

329 see :py:obj:`~.nvrtcGetSupportedArchs` 

330  

331 Returns 

332 ------- 

333 nvrtcResult 

334 - :py:obj:`~.NVRTC_SUCCESS` 

335 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

336 numArchs : int 

337 number of supported architectures. 

338 """ 

339 cdef int numArchs = 0 1 ctx1E

340 with nogil: 1 ctx1E

341 err = cynvrtc.nvrtcGetNumSupportedArchs(&numArchs) 1 ctx1E

342 if err != cynvrtc.NVRTC_SUCCESS: 1 ctx1E

343 return (_nvrtcResult(err), None) 

344 return (_nvrtcResult_SUCCESS, numArchs) 1 ctx1E

345  

346@cython.embedsignature(True) (empty)

347def nvrtcGetSupportedArchs(): 

348 """ nvrtcGetSupportedArchs populates the array passed via the output parameter `supportedArchs` with the architectures supported by NVRTC. The array is sorted in the ascending order. The size of the array to be passed can be determined using :py:obj:`~.nvrtcGetNumSupportedArchs`. 

349  

350 see :py:obj:`~.nvrtcGetNumSupportedArchs` 

351  

352 Returns 

353 ------- 

354 nvrtcResult 

355 - :py:obj:`~.NVRTC_SUCCESS` 

356 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

357 supportedArchs : list[int] 

358 sorted array of supported architectures. 

359 """ 

360 cdef vector[int] supportedArchs 

361 _, s = nvrtcGetNumSupportedArchs() 1 ctx1E

362 supportedArchs.resize(s) 2 ctx1aE

363  

364 with nogil: 1 ctx1E

365 err = cynvrtc.nvrtcGetSupportedArchs(supportedArchs.data()) 1 ctx1E

366 if err != cynvrtc.NVRTC_SUCCESS: 1 ctx1E

367 return (_nvrtcResult(err), None) 

368 return (_nvrtcResult_SUCCESS, supportedArchs) 1 ctx1E

369  

370@cython.embedsignature(True) (empty)

371def nvrtcCreateProgram(char* src, char* name, int numHeaders, headers : Optional[tuple[bytes] | list[bytes]], includeNames : Optional[tuple[bytes] | list[bytes]]): 

372 """ nvrtcCreateProgram creates an instance of :py:obj:`~.nvrtcProgram` with the given input parameters, and sets the output parameter `prog` with it. 

373  

374 Parameters 

375 ---------- 

376 src : bytes 

377 CUDA program source. 

378 name : bytes 

379 CUDA program name. `name` can be `NULL`; `"default_program"` is 

380 used when `name` is `NULL` or "". 

381 numHeaders : int 

382 Number of headers used. `numHeaders` must be greater than or equal 

383 to 0. 

384 headers : list[bytes] 

385 Sources of the headers. `headers` can be `NULL` when `numHeaders` 

386 is 0. 

387 includeNames : list[bytes] 

388 Name of each header by which they can be included in the CUDA 

389 program source. `includeNames` can be `NULL` when `numHeaders` is 

390 0. These headers must be included with the exact names specified 

391 here. 

392  

393 Returns 

394 ------- 

395 nvrtcResult 

396 - :py:obj:`~.NVRTC_SUCCESS` 

397 - :py:obj:`~.NVRTC_ERROR_OUT_OF_MEMORY` 

398 - :py:obj:`~.NVRTC_ERROR_PROGRAM_CREATION_FAILURE` 

399 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

400 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

401 prog : :py:obj:`~.nvrtcProgram` 

402 CUDA Runtime Compilation program. 

403  

404 See Also 

405 -------- 

406 :py:obj:`~.nvrtcDestroyProgram` 

407 """ 

408 includeNames = [] if includeNames is None else includeNames 27 ctx1bcdefghiklmnopqrstuvwxyzABj

409 if not all(isinstance(_x, (bytes)) for _x in includeNames): 27 ctx1bcdefghiklmnopqrstuvwxyzABj

410 raise TypeError("Argument 'includeNames' is not instance of type (expected tuple[bytes] or list[bytes]") 

411 headers = [] if headers is None else headers 27 ctx1bcdefghiklmnopqrstuvwxyzABj

412 if not all(isinstance(_x, (bytes)) for _x in headers): 27 ctx1bcdefghiklmnopqrstuvwxyzABj

413 raise TypeError("Argument 'headers' is not instance of type (expected tuple[bytes] or list[bytes]") 

414 cdef nvrtcProgram prog = nvrtcProgram() 27 ctx1bcdefghiklmnopqrstuvwxyzABj

415 if numHeaders > len(headers): raise RuntimeError("List is too small: " + str(len(headers)) + " < " + str(numHeaders)) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

416 if numHeaders > len(includeNames): raise RuntimeError("List is too small: " + str(len(includeNames)) + " < " + str(numHeaders)) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

417 cdef vector[const char*] cyheaders = headers 27 ctx1bcdefghiklmnopqrstuvwxyzABj

418 cdef vector[const char*] cyincludeNames = includeNames 27 ctx1bcdefghiklmnopqrstuvwxyzABj

419 with nogil: 27 ctx1bcdefghiklmnopqrstuvwxyzABj

420 err = cynvrtc.nvrtcCreateProgram(<cynvrtc.nvrtcProgram*>prog._pvt_ptr, src, name, numHeaders, cyheaders.data(), cyincludeNames.data()) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

421 if err != cynvrtc.NVRTC_SUCCESS: 27 ctx1bcdefghiklmnopqrstuvwxyzABj

422 return (_nvrtcResult(err), None) 

423 return (_nvrtcResult_SUCCESS, prog) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

424  

425@cython.embedsignature(True) (empty)

426def nvrtcDestroyProgram(prog): 

427 """ nvrtcDestroyProgram destroys the given program. 

428  

429 Parameters 

430 ---------- 

431 prog : :py:obj:`~.nvrtcProgram` 

432 CUDA Runtime Compilation program. 

433  

434 Returns 

435 ------- 

436 nvrtcResult 

437 - :py:obj:`~.NVRTC_SUCCESS` 

438 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

439  

440 See Also 

441 -------- 

442 :py:obj:`~.nvrtcCreateProgram` 

443 """ 

444 cdef cynvrtc.nvrtcProgram *cyprog 

445 if prog is None: 19 ctx1klmnopqrstuvwxyzABj

446 cyprog = <cynvrtc.nvrtcProgram*><void_ptr>NULL 

447 elif isinstance(prog, (nvrtcProgram,)): 19 ctx1klmnopqrstuvwxyzABj

448 pprog = prog.getPtr() 19 ctx1klmnopqrstuvwxyzABj

449 cyprog = <cynvrtc.nvrtcProgram*><void_ptr>pprog 19 ctx1klmnopqrstuvwxyzABj

450 elif isinstance(prog, (int)): 

451 cyprog = <cynvrtc.nvrtcProgram*><void_ptr>prog 

452 else: 

453 raise TypeError("Argument 'prog' is not instance of type (expected <class 'int, nvrtc.nvrtcProgram'>, found " + str(type(prog))) 

454 with nogil: 19 ctx1klmnopqrstuvwxyzABj

455 err = cynvrtc.nvrtcDestroyProgram(cyprog) 19 ctx1klmnopqrstuvwxyzABj

456 return (_nvrtcResult(err),) 19 ctx1klmnopqrstuvwxyzABj

457  

458@cython.embedsignature(True) (empty)

459def nvrtcCompileProgram(prog, int numOptions, options : Optional[tuple[bytes] | list[bytes]]): 

460 """ nvrtcCompileProgram compiles the given program. 

461  

462 It supports compile options listed in :py:obj:`~.Supported Compile 

463 Options`. 

464  

465 Parameters 

466 ---------- 

467 prog : :py:obj:`~.nvrtcProgram` 

468 CUDA Runtime Compilation program. 

469 numOptions : int 

470 Number of compiler options passed. 

471 options : list[bytes] 

472 Compiler options in the form of C string array. `options` can be 

473 `NULL` when `numOptions` is 0. 

474  

475 Returns 

476 ------- 

477 nvrtcResult 

478 - :py:obj:`~.NVRTC_SUCCESS` 

479 - :py:obj:`~.NVRTC_ERROR_OUT_OF_MEMORY` 

480 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

481 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

482 - :py:obj:`~.NVRTC_ERROR_INVALID_OPTION` 

483 - :py:obj:`~.NVRTC_ERROR_COMPILATION` 

484 - :py:obj:`~.NVRTC_ERROR_BUILTIN_OPERATION_FAILURE` 

485 - :py:obj:`~.NVRTC_ERROR_TIME_FILE_WRITE_FAILED` 

486 - :py:obj:`~.NVRTC_ERROR_CANCELLED` 

487 """ 

488 options = [] if options is None else options 27 ctx1bcdefghiklmnopqrstuvwxyzABj

489 if not all(isinstance(_x, (bytes)) for _x in options): 27 ctx1bcdefghiklmnopqrstuvwxyzABj

490 raise TypeError("Argument 'options' is not instance of type (expected tuple[bytes] or list[bytes]") 

491 cdef cynvrtc.nvrtcProgram cyprog 

492 if prog is None: 27 ctx1bcdefghiklmnopqrstuvwxyzABj

493 pprog = 0 

494 elif isinstance(prog, (nvrtcProgram,)): 27 ctx1bcdefghiklmnopqrstuvwxyzABj

495 pprog = int(prog) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

496 else: 

497 pprog = int(nvrtcProgram(prog)) 

498 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 27 ctx1bcdefghiklmnopqrstuvwxyzABj

499 if numOptions > len(options): raise RuntimeError("List is too small: " + str(len(options)) + " < " + str(numOptions)) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

500 cdef vector[const char*] cyoptions = options 27 ctx1bcdefghiklmnopqrstuvwxyzABj

501 with nogil: 27 ctx1bcdefghiklmnopqrstuvwxyzABj

502 err = cynvrtc.nvrtcCompileProgram(cyprog, numOptions, cyoptions.data()) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

503 return (_nvrtcResult(err),) 27 ctx1bcdefghiklmnopqrstuvwxyzABj

504  

505@cython.embedsignature(True) (empty)

506def nvrtcGetPTXSize(prog): 

507 """ nvrtcGetPTXSize sets the value of `ptxSizeRet` with the size of the PTX generated by the previous compilation of `prog` (including the trailing `NULL`). 

508  

509 Parameters 

510 ---------- 

511 prog : :py:obj:`~.nvrtcProgram` 

512 CUDA Runtime Compilation program. 

513  

514 Returns 

515 ------- 

516 nvrtcResult 

517 - :py:obj:`~.NVRTC_SUCCESS` 

518 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

519 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

520 ptxSizeRet : int 

521 Size of the generated PTX (including the trailing `NULL`). 

522  

523 See Also 

524 -------- 

525 :py:obj:`~.nvrtcGetPTX` 

526 """ 

527 cdef cynvrtc.nvrtcProgram cyprog 

528 if prog is None: 

529 pprog = 0 

530 elif isinstance(prog, (nvrtcProgram,)): 

531 pprog = int(prog) 

532 else: 

533 pprog = int(nvrtcProgram(prog)) 

534 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

535 cdef size_t ptxSizeRet = 0 

536 with nogil: 

537 err = cynvrtc.nvrtcGetPTXSize(cyprog, &ptxSizeRet) 

538 if err != cynvrtc.NVRTC_SUCCESS: 

539 return (_nvrtcResult(err), None) 

540 return (_nvrtcResult_SUCCESS, ptxSizeRet) 

541  

542@cython.embedsignature(True) (empty)

543def nvrtcGetPTX(prog, char* ptx): 

544 """ nvrtcGetPTX stores the PTX generated by the previous compilation of `prog` in the memory pointed by `ptx`. 

545  

546 Parameters 

547 ---------- 

548 prog : :py:obj:`~.nvrtcProgram` 

549 CUDA Runtime Compilation program. 

550 ptx : bytes 

551 Compiled result. 

552  

553 Returns 

554 ------- 

555 nvrtcResult 

556 - :py:obj:`~.NVRTC_SUCCESS` 

557 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

558 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

559  

560 See Also 

561 -------- 

562 :py:obj:`~.nvrtcGetPTXSize` 

563 """ 

564 cdef cynvrtc.nvrtcProgram cyprog 

565 if prog is None: 

566 pprog = 0 

567 elif isinstance(prog, (nvrtcProgram,)): 

568 pprog = int(prog) 

569 else: 

570 pprog = int(nvrtcProgram(prog)) 

571 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

572 with nogil: 

573 err = cynvrtc.nvrtcGetPTX(cyprog, ptx) 

574 return (_nvrtcResult(err),) (empty)

575  

576@cython.embedsignature(True) (empty)

577def nvrtcGetCUBINSize(prog): 

578 """ nvrtcGetCUBINSize sets the value of `cubinSizeRet` with the size of the cubin generated by the previous compilation of `prog`. The value of cubinSizeRet is set to 0 if the value specified to `-arch` is a virtual architecture instead of an actual architecture. 

579  

580 Parameters 

581 ---------- 

582 prog : :py:obj:`~.nvrtcProgram` 

583 CUDA Runtime Compilation program. 

584  

585 Returns 

586 ------- 

587 nvrtcResult 

588 - :py:obj:`~.NVRTC_SUCCESS` 

589 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

590 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

591 cubinSizeRet : int 

592 Size of the generated cubin. 

593  

594 See Also 

595 -------- 

596 :py:obj:`~.nvrtcGetCUBIN` 

597 """ 

598 cdef cynvrtc.nvrtcProgram cyprog 

599 if prog is None: 14 ctx1bcdefghiopqrst

600 pprog = 0 

601 elif isinstance(prog, (nvrtcProgram,)): 14 ctx1bcdefghiopqrst

602 pprog = int(prog) 14 ctx1bcdefghiopqrst

603 else: 

604 pprog = int(nvrtcProgram(prog)) 

605 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 14 ctx1bcdefghiopqrst

606 cdef size_t cubinSizeRet = 0 14 ctx1bcdefghiopqrst

607 with nogil: 14 ctx1bcdefghiopqrst

608 err = cynvrtc.nvrtcGetCUBINSize(cyprog, &cubinSizeRet) 14 ctx1bcdefghiopqrst

609 if err != cynvrtc.NVRTC_SUCCESS: 14 ctx1bcdefghiopqrst

610 return (_nvrtcResult(err), None) 

611 return (_nvrtcResult_SUCCESS, cubinSizeRet) 14 ctx1bcdefghiopqrst

612  

613@cython.embedsignature(True) (empty)

614def nvrtcGetCUBIN(prog, char* cubin): 

615 """ nvrtcGetCUBIN stores the cubin generated by the previous compilation of `prog` in the memory pointed by `cubin`. No cubin is available if the value specified to `-arch` is a virtual architecture instead of an actual architecture. The cubin does not contain code for the Tile functions (`__tile__` / `__tile_global__`) or variables (`__tile__`); use `nvrtcGetTileIR()` to extract the cuda_tile IR generated for Tile code. 

616  

617 Parameters 

618 ---------- 

619 prog : :py:obj:`~.nvrtcProgram` 

620 CUDA Runtime Compilation program. 

621 cubin : bytes 

622 Compiled and assembled result. 

623  

624 Returns 

625 ------- 

626 nvrtcResult 

627 - :py:obj:`~.NVRTC_SUCCESS` 

628 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

629 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

630  

631 See Also 

632 -------- 

633 :py:obj:`~.nvrtcGetCUBINSize` 

634 """ 

635 cdef cynvrtc.nvrtcProgram cyprog 

636 if prog is None: 14 ctx1bcdefghiopqrst

637 pprog = 0 

638 elif isinstance(prog, (nvrtcProgram,)): 14 ctx1bcdefghiopqrst

639 pprog = int(prog) 14 ctx1bcdefghiopqrst

640 else: 

641 pprog = int(nvrtcProgram(prog)) 

642 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 14 ctx1bcdefghiopqrst

643 with nogil: 14 ctx1bcdefghiopqrst

644 err = cynvrtc.nvrtcGetCUBIN(cyprog, cubin) 14 ctx1bcdefghiopqrst

645 return (_nvrtcResult(err),) 14 ctx1bcdefghiopqrst

646  

647@cython.embedsignature(True) (empty)

648def nvrtcGetLTOIRSize(prog): 

649 """ nvrtcGetLTOIRSize sets the value of `LTOIRSizeRet` with the size of the LTO IR generated by the previous compilation of `prog`. The value of LTOIRSizeRet is set to 0 if the program was not compiled with `-dlto`. 

650  

651 Parameters 

652 ---------- 

653 prog : :py:obj:`~.nvrtcProgram` 

654 CUDA Runtime Compilation program. 

655  

656 Returns 

657 ------- 

658 nvrtcResult 

659 - :py:obj:`~.NVRTC_SUCCESS` 

660 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

661 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

662 LTOIRSizeRet : int 

663 Size of the generated LTO IR. 

664  

665 See Also 

666 -------- 

667 :py:obj:`~.nvrtcGetLTOIR` 

668 """ 

669 cdef cynvrtc.nvrtcProgram cyprog 

670 if prog is None: 12 ctx1klmnuvwxyzAB

671 pprog = 0 

672 elif isinstance(prog, (nvrtcProgram,)): 12 ctx1klmnuvwxyzAB

673 pprog = int(prog) 12 ctx1klmnuvwxyzAB

674 else: 

675 pprog = int(nvrtcProgram(prog)) 

676 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 12 ctx1klmnuvwxyzAB

677 cdef size_t LTOIRSizeRet = 0 12 ctx1klmnuvwxyzAB

678 with nogil: 12 ctx1klmnuvwxyzAB

679 err = cynvrtc.nvrtcGetLTOIRSize(cyprog, &LTOIRSizeRet) 12 ctx1klmnuvwxyzAB

680 if err != cynvrtc.NVRTC_SUCCESS: 12 ctx1klmnuvwxyzAB

681 return (_nvrtcResult(err), None) 

682 return (_nvrtcResult_SUCCESS, LTOIRSizeRet) 12 ctx1klmnuvwxyzAB

683  

684@cython.embedsignature(True) (empty)

685def nvrtcGetLTOIR(prog, char* LTOIR): 

686 """ nvrtcGetLTOIR stores the LTO IR generated by the previous compilation of `prog` in the memory pointed by `LTOIR`. No LTO IR is available if the program was compiled without `-dlto`. 

687  

688 Parameters 

689 ---------- 

690 prog : :py:obj:`~.nvrtcProgram` 

691 CUDA Runtime Compilation program. 

692 LTOIR : bytes 

693 Compiled result. 

694  

695 Returns 

696 ------- 

697 nvrtcResult 

698 - :py:obj:`~.NVRTC_SUCCESS` 

699 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

700 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

701  

702 See Also 

703 -------- 

704 :py:obj:`~.nvrtcGetLTOIRSize` 

705 """ 

706 cdef cynvrtc.nvrtcProgram cyprog 

707 if prog is None: 12 ctx1klmnuvwxyzAB

708 pprog = 0 

709 elif isinstance(prog, (nvrtcProgram,)): 12 ctx1klmnuvwxyzAB

710 pprog = int(prog) 12 ctx1klmnuvwxyzAB

711 else: 

712 pprog = int(nvrtcProgram(prog)) 

713 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 12 ctx1klmnuvwxyzAB

714 with nogil: 12 ctx1klmnuvwxyzAB

715 err = cynvrtc.nvrtcGetLTOIR(cyprog, LTOIR) 12 ctx1klmnuvwxyzAB

716 return (_nvrtcResult(err),) 12 ctx1klmnuvwxyzAB

717  

718@cython.embedsignature(True) (empty)

719def nvrtcGetOptiXIRSize(prog): 

720 """ nvrtcGetOptiXIRSize sets the value of `optixirSizeRet` with the size of the OptiX IR generated by the previous compilation of `prog`. The value of nvrtcGetOptiXIRSize is set to 0 if the program was compiled with options incompatible with OptiX IR generation. 

721  

722 Parameters 

723 ---------- 

724 prog : :py:obj:`~.nvrtcProgram` 

725 CUDA Runtime Compilation program. 

726  

727 Returns 

728 ------- 

729 nvrtcResult 

730 - :py:obj:`~.NVRTC_SUCCESS` 

731 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

732 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

733 optixirSizeRet : int 

734 Size of the generated LTO IR. 

735  

736 See Also 

737 -------- 

738 :py:obj:`~.nvrtcGetOptiXIR` 

739 """ 

740 cdef cynvrtc.nvrtcProgram cyprog 

741 if prog is None: 

742 pprog = 0 

743 elif isinstance(prog, (nvrtcProgram,)): 

744 pprog = int(prog) 

745 else: 

746 pprog = int(nvrtcProgram(prog)) 

747 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

748 cdef size_t optixirSizeRet = 0 

749 with nogil: 

750 err = cynvrtc.nvrtcGetOptiXIRSize(cyprog, &optixirSizeRet) 

751 if err != cynvrtc.NVRTC_SUCCESS: 

752 return (_nvrtcResult(err), None) 

753 return (_nvrtcResult_SUCCESS, optixirSizeRet) 

754  

755@cython.embedsignature(True) (empty)

756def nvrtcGetOptiXIR(prog, char* optixir): 

757 """ nvrtcGetOptiXIR stores the OptiX IR generated by the previous compilation of `prog` in the memory pointed by `optixir`. No OptiX IR is available if the program was compiled with options incompatible with OptiX IR generation. 

758  

759 Parameters 

760 ---------- 

761 prog : :py:obj:`~.nvrtcProgram` 

762 CUDA Runtime Compilation program. 

763 optixir : bytes 

764 Optix IR Compiled result. 

765  

766 Returns 

767 ------- 

768 nvrtcResult 

769 - :py:obj:`~.NVRTC_SUCCESS` 

770 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

771 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

772  

773 See Also 

774 -------- 

775 :py:obj:`~.nvrtcGetOptiXIRSize` 

776 """ 

777 cdef cynvrtc.nvrtcProgram cyprog 

778 if prog is None: 

779 pprog = 0 

780 elif isinstance(prog, (nvrtcProgram,)): 

781 pprog = int(prog) 

782 else: 

783 pprog = int(nvrtcProgram(prog)) 

784 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

785 with nogil: 

786 err = cynvrtc.nvrtcGetOptiXIR(cyprog, optixir) 

787 return (_nvrtcResult(err),) 

788  

789@cython.embedsignature(True) (empty)

790def nvrtcGetProgramLogSize(prog): 

791 """ nvrtcGetProgramLogSize sets `logSizeRet` with the size of the log generated by the previous compilation of `prog` (including the trailing `NULL`). 

792  

793 Note that compilation log may be generated with warnings and 

794 informative messages, even when the compilation of `prog` succeeds. 

795  

796 Parameters 

797 ---------- 

798 prog : :py:obj:`~.nvrtcProgram` 

799 CUDA Runtime Compilation program. 

800  

801 Returns 

802 ------- 

803 nvrtcResult 

804 - :py:obj:`~.NVRTC_SUCCESS` 

805 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

806 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

807 logSizeRet : int 

808 Size of the compilation log (including the trailing `NULL`). 

809  

810 See Also 

811 -------- 

812 :py:obj:`~.nvrtcGetProgramLog` 

813 """ 

814 cdef cynvrtc.nvrtcProgram cyprog 

815 if prog is None: 9 ctx1bcdefghij

816 pprog = 0 

817 elif isinstance(prog, (nvrtcProgram,)): 9 ctx1bcdefghij

818 pprog = int(prog) 9 ctx1bcdefghij

819 else: 

820 pprog = int(nvrtcProgram(prog)) 

821 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 9 ctx1bcdefghij

822 cdef size_t logSizeRet = 0 9 ctx1bcdefghij

823 with nogil: 9 ctx1bcdefghij

824 err = cynvrtc.nvrtcGetProgramLogSize(cyprog, &logSizeRet) 9 ctx1bcdefghij

825 if err != cynvrtc.NVRTC_SUCCESS: 9 ctx1bcdefghij

826 return (_nvrtcResult(err), None) 

827 return (_nvrtcResult_SUCCESS, logSizeRet) 9 ctx1bcdefghij

828  

829@cython.embedsignature(True) (empty)

830def nvrtcGetProgramLog(prog, char* log): 

831 """ nvrtcGetProgramLog stores the log generated by the previous compilation of `prog` in the memory pointed by `log`. 

832  

833 Parameters 

834 ---------- 

835 prog : :py:obj:`~.nvrtcProgram` 

836 CUDA Runtime Compilation program. 

837 log : bytes 

838 Compilation log. 

839  

840 Returns 

841 ------- 

842 nvrtcResult 

843 - :py:obj:`~.NVRTC_SUCCESS` 

844 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

845 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

846  

847 See Also 

848 -------- 

849 :py:obj:`~.nvrtcGetProgramLogSize` 

850 """ 

851 cdef cynvrtc.nvrtcProgram cyprog 

852 if prog is None: 9 ctx1bcdefghij

853 pprog = 0 

854 elif isinstance(prog, (nvrtcProgram,)): 9 ctx1bcdefghij

855 pprog = int(prog) 9 ctx1bcdefghij

856 else: 

857 pprog = int(nvrtcProgram(prog)) 

858 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 9 ctx1bcdefghij

859 with nogil: 9 ctx1bcdefghij

860 err = cynvrtc.nvrtcGetProgramLog(cyprog, log) 9 ctx1bcdefghij

861 return (_nvrtcResult(err),) 9 ctx1bcdefghij

862  

863@cython.embedsignature(True) (empty)

864def nvrtcAddNameExpression(prog, char* name_expression): 

865 """ nvrtcAddNameExpression notes the given name expression denoting the address of a global function or device/__constant__ variable. 

866  

867 The identical name expression string must be provided on a subsequent 

868 call to nvrtcGetLoweredName to extract the lowered name. 

869  

870 Parameters 

871 ---------- 

872 prog : :py:obj:`~.nvrtcProgram` 

873 CUDA Runtime Compilation program. 

874 name_expression : bytes 

875 constant expression denoting the address of a global function or 

876 device/__constant__ variable. 

877  

878 Returns 

879 ------- 

880 nvrtcResult 

881 - :py:obj:`~.NVRTC_SUCCESS` 

882 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

883 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

884 - :py:obj:`~.NVRTC_ERROR_NO_NAME_EXPRESSIONS_AFTER_COMPILATION` 

885  

886 See Also 

887 -------- 

888 :py:obj:`~.nvrtcGetLoweredName` 

889 """ 

890 cdef cynvrtc.nvrtcProgram cyprog 

891 if prog is None: 

892 pprog = 0 

893 elif isinstance(prog, (nvrtcProgram,)): 

894 pprog = int(prog) 

895 else: 

896 pprog = int(nvrtcProgram(prog)) 

897 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

898 with nogil: 

899 err = cynvrtc.nvrtcAddNameExpression(cyprog, name_expression) 

900 return (_nvrtcResult(err),) 

901  

902@cython.embedsignature(True) (empty)

903def nvrtcGetLoweredName(prog, char* name_expression): 

904 """ nvrtcGetLoweredName extracts the lowered (mangled) name for a global function or device/__constant__ variable, and updates lowered_name to point to it. The memory containing the name is released when the NVRTC program is destroyed by nvrtcDestroyProgram. The identical name expression must have been previously provided to nvrtcAddNameExpression. 

905  

906 Parameters 

907 ---------- 

908 prog : nvrtcProgram 

909 CUDA Runtime Compilation program. 

910 name_expression : bytes 

911 constant expression denoting the address of a global function or 

912 device/__constant__ variable. 

913  

914 Returns 

915 ------- 

916 nvrtcResult 

917 NVRTC_SUCCESS 

918 NVRTC_ERROR_NO_LOWERED_NAMES_BEFORE_COMPILATION 

919 NVRTC_ERROR_NAME_EXPRESSION_NOT_VALID 

920 lowered_name : bytes 

921 initialized by the function to point to a C string containing the 

922 lowered (mangled) name corresponding to the provided name 

923 expression. 

924  

925 See Also 

926 -------- 

927 nvrtcAddNameExpression 

928 """ 

929 cdef cynvrtc.nvrtcProgram cyprog 

930 if prog is None: 1 ctx1D

931 pprog = 0 1 ctx1D

932 elif isinstance(prog, (nvrtcProgram,)): 1 ctx1D

933 pprog = int(prog) 

934 else: 

935 pprog = int(nvrtcProgram(prog)) 1 ctx1D

936 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 1 ctx1D

937 cdef const char* lowered_name = NULL 1 ctx1D

938 with nogil: 1 ctx1D

939 err = cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name) 1 ctx1D

940 if err != cynvrtc.NVRTC_SUCCESS: 1 ctx1D

941 return (_nvrtcResult(err), None) 1 ctx1D

942 return (_nvrtcResult_SUCCESS, <bytes>lowered_name if lowered_name != NULL else None) 

943  

944@cython.embedsignature(True) (empty)

945def nvrtcGetPCHHeapSize(): 

946 """ retrieve the current size of the PCH Heap. 

947  

948 Returns 

949 ------- 

950 nvrtcResult 

951 - :py:obj:`~.NVRTC_SUCCESS` 

952 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

953 ret : int 

954 pointer to location where the size of the PCH Heap will be stored 

955 """ 

956 cdef size_t ret = 0 

957 with nogil: 

958 err = cynvrtc.nvrtcGetPCHHeapSize(&ret) 

959 if err != cynvrtc.NVRTC_SUCCESS: 

960 return (_nvrtcResult(err), None) 

961 return (_nvrtcResult_SUCCESS, ret) 

962  

963@cython.embedsignature(True) (empty)

964def nvrtcSetPCHHeapSize(size_t size): 

965 """ set the size of the PCH Heap. 

966  

967 The requested size may be rounded up to a platform dependent alignment 

968 (e.g. page size). If the PCH Heap has already been allocated, the heap 

969 memory will be freed and a new PCH Heap will be allocated. 

970  

971 Parameters 

972 ---------- 

973 size : size_t 

974 requested size of the PCH Heap, in bytes 

975  

976 Returns 

977 ------- 

978 nvrtcResult 

979 - :py:obj:`~.NVRTC_SUCCESS` 

980 """ 

981 with nogil: 

982 err = cynvrtc.nvrtcSetPCHHeapSize(size) 

983 return (_nvrtcResult(err),) 

984  

985@cython.embedsignature(True) (empty)

986def nvrtcGetPCHCreateStatus(prog): 

987 """ returns the PCH creation status. 

988  

989 NVRTC_SUCCESS indicates that the PCH was successfully created. 

990 NVRTC_ERROR_NO_PCH_CREATE_ATTEMPTED indicates that no PCH creation was 

991 attempted, either because PCH functionality was not requested during 

992 the preceding nvrtcCompileProgram call, or automatic PCH processing was 

993 requested, and compiler chose not to create a PCH file. 

994 NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED indicates that a PCH file could 

995 potentially have been created, but the compiler ran out space in the 

996 PCH heap. In this scenario, the 

997 :py:obj:`~.nvrtcGetPCHHeapSizeRequired()` can be used to query the 

998 required heap size, the heap can be reallocated for this size with 

999 :py:obj:`~.nvrtcSetPCHHeapSize()` and PCH creation may be reattempted 

1000 again invoking :py:obj:`~.nvrtcCompileProgram()` with a new NVRTC 

1001 program instance. NVRTC_ERROR_PCH_CREATE indicates that an error 

1002 condition prevented the PCH file from being created. 

1003  

1004 Parameters 

1005 ---------- 

1006 prog : :py:obj:`~.nvrtcProgram` 

1007 CUDA Runtime Compilation program. 

1008  

1009 Returns 

1010 ------- 

1011 nvrtcResult 

1012 - :py:obj:`~.NVRTC_SUCCESS` 

1013 - :py:obj:`~.NVRTC_ERROR_NO_PCH_CREATE_ATTEMPTED` 

1014 - :py:obj:`~.NVRTC_ERROR_PCH_CREATE` 

1015 - :py:obj:`~.NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED` 

1016 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

1017 """ 

1018 cdef cynvrtc.nvrtcProgram cyprog 

1019 if prog is None: 

1020 pprog = 0 

1021 elif isinstance(prog, (nvrtcProgram,)): 

1022 pprog = int(prog) 

1023 else: 

1024 pprog = int(nvrtcProgram(prog)) 

1025 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

1026 with nogil: 

1027 err = cynvrtc.nvrtcGetPCHCreateStatus(cyprog) 

1028 return (_nvrtcResult(err),) 

1029  

1030@cython.embedsignature(True) (empty)

1031def nvrtcGetPCHHeapSizeRequired(prog): 

1032 """ retrieve the required size of the PCH heap required to compile the given program. 

1033  

1034 Parameters 

1035 ---------- 

1036 prog : :py:obj:`~.nvrtcProgram` 

1037 CUDA Runtime Compilation program. 

1038  

1039 Returns 

1040 ------- 

1041 nvrtcResult 

1042 - :py:obj:`~.NVRTC_SUCCESS` 

1043 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

1044 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` The size retrieved using this function is only valid if :py:obj:`~.nvrtcGetPCHCreateStatus()` returned NVRTC_SUCCESS or NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED 

1045 size : int 

1046 pointer to location where the required size of the PCH Heap will be 

1047 stored 

1048 """ 

1049 cdef cynvrtc.nvrtcProgram cyprog 

1050 if prog is None: 

1051 pprog = 0 

1052 elif isinstance(prog, (nvrtcProgram,)): 

1053 pprog = int(prog) 

1054 else: 

1055 pprog = int(nvrtcProgram(prog)) 

1056 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

1057 cdef size_t size = 0 

1058 with nogil: 

1059 err = cynvrtc.nvrtcGetPCHHeapSizeRequired(cyprog, &size) 

1060 if err != cynvrtc.NVRTC_SUCCESS: 

1061 return (_nvrtcResult(err), None) 

1062 return (_nvrtcResult_SUCCESS, size) 

1063  

1064@cython.embedsignature(True) (empty)

1065def nvrtcSetFlowCallback(prog, callback, payload): 

1066 """ nvrtcSetFlowCallback registers a callback function that the compiler will invoke at different points during a call to nvrtcCompileProgram, and the callback function can decide whether to cancel compilation by returning specific values. 

1067  

1068 The callback function must satisfy the following constraints: 

1069  

1070 (1) Its signature should be: 

1071  

1072 **View CUDA Toolkit Documentation for a C++ code example** 

1073  

1074 When invoking the callback, the compiler will always pass `payload` to 

1075 param1 so that the callback may make decisions based on `payload` . 

1076 It'll always pass NULL to param2 for now which is reserved for future 

1077 extensions. 

1078  

1079 (2) It must return 1 to cancel compilation or 0 to continue. Other 

1080 return values are reserved for future use. 

1081  

1082 (3) It must return consistent values. Once it returns 1 at one point, 

1083 it must return 1 in all following invocations during the current 

1084 nvrtcCompileProgram call in progress. 

1085  

1086 (4) It must be thread-safe. 

1087  

1088 (5) It must not invoke any nvrtc/libnvvm/ptx APIs. 

1089  

1090 Parameters 

1091 ---------- 

1092 prog : :py:obj:`~.nvrtcProgram` 

1093 CUDA Runtime Compilation program. 

1094 callback : Any 

1095 the callback that issues cancellation signal. 

1096 payload : Any 

1097 to be passed as a parameter when invoking the callback. 

1098  

1099 Returns 

1100 ------- 

1101 nvrtcResult 

1102 - :py:obj:`~.NVRTC_SUCCESS` 

1103 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

1104 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

1105 """ 

1106 cdef cynvrtc.nvrtcProgram cyprog 

1107 if prog is None: 

1108 pprog = 0 

1109 elif isinstance(prog, (nvrtcProgram,)): 

1110 pprog = int(prog) 

1111 else: 

1112 pprog = int(nvrtcProgram(prog)) 

1113 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

1114 cdef _HelperInputVoidPtrStruct cycallbackHelper 

1115 cdef void* cycallback = _helper_input_void_ptr(callback, &cycallbackHelper) 

1116 cdef _HelperInputVoidPtrStruct cypayloadHelper 

1117 cdef void* cypayload = _helper_input_void_ptr(payload, &cypayloadHelper) 

1118 with nogil: 

1119 err = cynvrtc.nvrtcSetFlowCallback(cyprog, cycallback, cypayload) 

1120 _helper_input_void_ptr_free(&cycallbackHelper) 

1121 _helper_input_void_ptr_free(&cypayloadHelper) 

1122 return (_nvrtcResult(err),) 

1123  

1124@cython.embedsignature(True) (empty)

1125def nvrtcGetTileIRSize(prog): 

1126 """ nvrtcGetTileIRSize sets the value of `TileIRSizeRet` with the size of the cuda_tile IR generated by the previous compilation of `prog`. 

1127  

1128 Parameters 

1129 ---------- 

1130 prog : :py:obj:`~.nvrtcProgram` 

1131 CUDA Runtime Compilation program. 

1132  

1133 Returns 

1134 ------- 

1135 nvrtcResult 

1136 - :py:obj:`~.NVRTC_SUCCESS` 

1137 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

1138 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

1139 TileIRSizeRet : int 

1140 Size of the generated cuda_tile IR. 

1141  

1142 See Also 

1143 -------- 

1144 :py:obj:`~.nvrtcGetTileIR` 

1145 """ 

1146 cdef cynvrtc.nvrtcProgram cyprog 

1147 if prog is None: 

1148 pprog = 0 

1149 elif isinstance(prog, (nvrtcProgram,)): 

1150 pprog = int(prog) 

1151 else: 

1152 pprog = int(nvrtcProgram(prog)) 

1153 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

1154 cdef size_t TileIRSizeRet = 0 

1155 with nogil: 

1156 err = cynvrtc.nvrtcGetTileIRSize(cyprog, &TileIRSizeRet) 

1157 if err != cynvrtc.NVRTC_SUCCESS: 

1158 return (_nvrtcResult(err), None) 

1159 return (_nvrtcResult_SUCCESS, TileIRSizeRet) 

1160  

1161@cython.embedsignature(True) (empty)

1162def nvrtcGetTileIR(prog, char* TileIR): 

1163 """ nvrtcGetTileIR stores the cuda_tile IR generated by the previous compilation of `prog` in the memory pointed by `TileIR`. 

1164  

1165 Parameters 

1166 ---------- 

1167 prog : :py:obj:`~.nvrtcProgram` 

1168 CUDA Runtime Compilation program. 

1169 TileIR : bytes 

1170 Generated cuda_tile IR. 

1171  

1172 Returns 

1173 ------- 

1174 nvrtcResult 

1175 - :py:obj:`~.NVRTC_SUCCESS` 

1176 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` 

1177 - :py:obj:`~.NVRTC_ERROR_INVALID_PROGRAM` 

1178  

1179 See Also 

1180 -------- 

1181 :py:obj:`~.nvrtcGetTileIRSize` 

1182 """ 

1183 cdef cynvrtc.nvrtcProgram cyprog 

1184 if prog is None: 

1185 pprog = 0 

1186 elif isinstance(prog, (nvrtcProgram,)): 

1187 pprog = int(prog) 

1188 else: 

1189 pprog = int(nvrtcProgram(prog)) 

1190 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 

1191 with nogil: 

1192 err = cynvrtc.nvrtcGetTileIR(cyprog, TileIR) 

1193 return (_nvrtcResult(err),) 

1194  

1195@cython.embedsignature(True) (empty)

1196def nvrtcInstallBundledHeaders(char* installPath, unsigned int flags): 

1197 """ nvrtcInstallBundledHeaders extracts CUDA headers bundled with NVRTC to a specified directory for use during compilation. 

1198  

1199 NVRTC bundles a set of CUDA Toolkit headers and CUDA C++ Core Libraries 

1200 (CCCL) within libnvrtc-builtins. This function extracts these headers 

1201 to the specified directory, allowing NVRTC programs to compile without 

1202 requiring a separate CUDA Toolkit installation. The bundled headers 

1203 match those available in the CUDA Toolkit plus CCCL libraries. 

1204  

1205 After extraction, users can compile kernels by passing appropriate 

1206 include paths (such as "-I<installPath>" and "-I<installPath>/cccl") to 

1207 nvrtcCompileProgram. 

1208  

1209 A version marker file (.nvrtc_headers_version) is created in the 

1210 installation directory to track the installed version. 

1211  

1212 This function is thread-safe and process-safe. Concurrent calls from 

1213 multiple threads or processes will be serialized using file locking. By 

1214 default, the function waits for the lock; use 

1215 NVRTC_INSTALL_HEADERS_NO_WAIT to return immediately with 

1216 NVRTC_ERROR_BUSY if another process holds the lock. 

1217  

1218 Parameters 

1219 ---------- 

1220 installPath : bytes 

1221 Path where headers should be extracted (UTF-8 encoded). The 

1222 directory will be created if it doesn't exist. 

1223 flags : unsigned int 

1224 NVRTC_INSTALL_HEADERS_* flags: 

1225  

1226 Returns 

1227 ------- 

1228 nvrtcResult 

1229 - :py:obj:`~.NVRTC_SUCCESS` 

1230 - :py:obj:`~.NVRTC_ERROR_INVALID_INPUT` (invalid path or conflicting flags like SKIP_IF_EXISTS | FORCE_OVERWRITE) 

1231 - :py:obj:`~.NVRTC_ERROR_BUILTIN_OPERATION_FAILURE` (extraction failed or version mismatch) 

1232 - :py:obj:`~.NVRTC_ERROR_BUSY` (lock held by another process and NVRTC_INSTALL_HEADERS_NO_WAIT was specified) 

1233 errorLog : bytes 

1234 Optional pointer to receive detailed error message on failure. If 

1235 non-NULL, `*errorLog` will be set to point to a string describing 

1236 the error cause. Note: subsequent API calls from the same thread 

1237 may overwrite this message. May be NULL if error details are not 

1238 needed. 

1239  

1240 See Also 

1241 -------- 

1242 :py:obj:`~.nvrtcCompileProgram` 

1243  

1244 Notes 

1245 ----- 

1246 Use NVRTC_INSTALL_HEADERS_SKIP_IF_EXISTS to avoid reinstalling if headers already exist. Use NVRTC_INSTALL_HEADERS_FORCE_OVERWRITE to guarantee consistency by clearing the directory first. 

1247 """ 

1248 cdef const char* errorLog = NULL 

1249 with nogil: 

1250 err = cynvrtc.nvrtcInstallBundledHeaders(installPath, flags, &errorLog) 

1251 if err != cynvrtc.NVRTC_SUCCESS: 

1252 return (_nvrtcResult(err), None) 

1253 return (_nvrtcResult_SUCCESS, <bytes>errorLog if errorLog != NULL else None) 

1254  

1255@cython.embedsignature(True) (empty)

1256def nvrtcGetBundledHeadersInfo(): 

1257 """ nvrtcGetBundledHeadersInfo queries information about the bundled headers without extracting them. 

1258  

1259 This function allows users to determine if bundled headers are 

1260 available and get size estimates before calling