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

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

2# SPDX-License-Identifier: LicenseRef-NVIDIA-SOFTWARE-LICENSE 

3  

4# This code was automatically generated with version 13.3.0, generator version 0.3.1.dev1738+g1060a290f. Do not modify it directly. 

5from typing import Any, Optional (empty)

6import cython 

7import ctypes (empty)

8from libc.stdlib cimport calloc, malloc, free 

9from libc cimport string 

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

11from libc.stddef cimport wchar_t 

12from libc.limits cimport CHAR_MIN 

13from libcpp.vector cimport vector 

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

15from cpython.bytes cimport PyBytes_FromStringAndSize 

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

17  

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

19_driver = _driver.__dict__ (empty)

20include "_lib/utils.pxi" 

21  

22ctypedef unsigned long long signed_char_ptr 

23ctypedef unsigned long long unsigned_char_ptr 

24ctypedef unsigned long long char_ptr 

25ctypedef unsigned long long short_ptr 

26ctypedef unsigned long long unsigned_short_ptr 

27ctypedef unsigned long long int_ptr 

28ctypedef unsigned long long long_int_ptr 

29ctypedef unsigned long long long_long_int_ptr 

30ctypedef unsigned long long unsigned_int_ptr 

31ctypedef unsigned long long unsigned_long_int_ptr 

32ctypedef unsigned long long unsigned_long_long_int_ptr 

33ctypedef unsigned long long uint32_t_ptr 

34ctypedef unsigned long long uint64_t_ptr 

35ctypedef unsigned long long int32_t_ptr 

36ctypedef unsigned long long int64_t_ptr 

37ctypedef unsigned long long unsigned_ptr 

38ctypedef unsigned long long unsigned_long_long_ptr 

39ctypedef unsigned long long long_long_ptr 

40ctypedef unsigned long long size_t_ptr 

41ctypedef unsigned long long long_ptr 

42ctypedef unsigned long long float_ptr 

43ctypedef unsigned long long double_ptr 

44ctypedef unsigned long long void_ptr 

45  

46#: Flags for nvrtcInstallBundledHeaders.Skip installation if version marker 

47#: exists and version matches. This is the default behavior when flags=0. 

48NVRTC_INSTALL_HEADERS_SKIP_IF_EXISTS = cynvrtc.NVRTC_INSTALL_HEADERS_SKIP_IF_EXISTS (empty)

49  

50#: Clear existing directory contents before installation. Guarantees 

51#: consistency by removing any existing files first. 

52NVRTC_INSTALL_HEADERS_FORCE_OVERWRITE = cynvrtc.NVRTC_INSTALL_HEADERS_FORCE_OVERWRITE (empty)

53  

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

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

56#: FORCE_OVERWRITE using bitwise OR. 

57NVRTC_INSTALL_HEADERS_NO_WAIT = cynvrtc.NVRTC_INSTALL_HEADERS_NO_WAIT (empty)

58  

59class nvrtcResult(_FastEnum): (empty)

60 """ 

61 The enumerated type nvrtcResult defines API call result codes. 

62 NVRTC API functions return nvrtcResult to indicate the call result. 

63 """ 

64  

65 NVRTC_SUCCESS = cynvrtc.nvrtcResult.NVRTC_SUCCESS (empty)

66  

67 NVRTC_ERROR_OUT_OF_MEMORY = cynvrtc.nvrtcResult.NVRTC_ERROR_OUT_OF_MEMORY (empty)

68  

69 NVRTC_ERROR_PROGRAM_CREATION_FAILURE = cynvrtc.nvrtcResult.NVRTC_ERROR_PROGRAM_CREATION_FAILURE (empty)

70  

71 NVRTC_ERROR_INVALID_INPUT = cynvrtc.nvrtcResult.NVRTC_ERROR_INVALID_INPUT (empty)

72  

73 NVRTC_ERROR_INVALID_PROGRAM = cynvrtc.nvrtcResult.NVRTC_ERROR_INVALID_PROGRAM (empty)

74  

75 NVRTC_ERROR_INVALID_OPTION = cynvrtc.nvrtcResult.NVRTC_ERROR_INVALID_OPTION (empty)

76  

77 NVRTC_ERROR_COMPILATION = cynvrtc.nvrtcResult.NVRTC_ERROR_COMPILATION (empty)

78  

79 NVRTC_ERROR_BUILTIN_OPERATION_FAILURE = cynvrtc.nvrtcResult.NVRTC_ERROR_BUILTIN_OPERATION_FAILURE (empty)

80  

81 NVRTC_ERROR_NO_NAME_EXPRESSIONS_AFTER_COMPILATION = cynvrtc.nvrtcResult.NVRTC_ERROR_NO_NAME_EXPRESSIONS_AFTER_COMPILATION (empty)

82  

83 NVRTC_ERROR_NO_LOWERED_NAMES_BEFORE_COMPILATION = cynvrtc.nvrtcResult.NVRTC_ERROR_NO_LOWERED_NAMES_BEFORE_COMPILATION (empty)

84  

85 NVRTC_ERROR_NAME_EXPRESSION_NOT_VALID = cynvrtc.nvrtcResult.NVRTC_ERROR_NAME_EXPRESSION_NOT_VALID (empty)

86  

87 NVRTC_ERROR_INTERNAL_ERROR = cynvrtc.nvrtcResult.NVRTC_ERROR_INTERNAL_ERROR (empty)

88  

89 NVRTC_ERROR_TIME_FILE_WRITE_FAILED = cynvrtc.nvrtcResult.NVRTC_ERROR_TIME_FILE_WRITE_FAILED (empty)

90  

91 NVRTC_ERROR_NO_PCH_CREATE_ATTEMPTED = cynvrtc.nvrtcResult.NVRTC_ERROR_NO_PCH_CREATE_ATTEMPTED (empty)

92  

93 NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED = cynvrtc.nvrtcResult.NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED (empty)

94  

95 NVRTC_ERROR_PCH_CREATE = cynvrtc.nvrtcResult.NVRTC_ERROR_PCH_CREATE (empty)

96  

97 NVRTC_ERROR_CANCELLED = cynvrtc.nvrtcResult.NVRTC_ERROR_CANCELLED (empty)

98  

99 NVRTC_ERROR_TIME_TRACE_FILE_WRITE_FAILED = cynvrtc.nvrtcResult.NVRTC_ERROR_TIME_TRACE_FILE_WRITE_FAILED (empty)

100  

101 NVRTC_ERROR_BUSY = cynvrtc.nvrtcResult.NVRTC_ERROR_BUSY (empty)

102  

103cdef object _nvrtcResult = nvrtcResult (empty)

104cdef object _nvrtcResult_SUCCESS = nvrtcResult.NVRTC_SUCCESS (empty)

105  

106cdef class nvrtcProgram: 

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

108  

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

110  

111 Methods 

112 ------- 

113 getPtr() 

114 Get memory address of class instance 

115  

116 """ 

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

118 if _ptr == 0: 26 ctx2j b c d e f g h i k l m n o p q r s t u v w x y ,b-b

119 self._pvt_ptr = &self._pvt_val 26 ctx2j b c d e f g h i k l m n o p q r s t u v w x y ,b-b

120 self._pvt_ptr[0] = <cynvrtc.nvrtcProgram>init_value 26 ctx2j b c d e f g h i k l m n o p q r s t u v w x y ,b-b

121 else: 

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

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

124 pass 26 ctx2j b c d e f g h i k l m n o p q r s t u v w x y ,b-b

125 def __repr__(self): 

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

127 def __index__(self): 

128 return self.__int__() 

129 def __eq__(self, other): 

130 if not isinstance(other, nvrtcProgram): 

131 return False 

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

133 def __hash__(self): 

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

135 def __int__(self): 

136 return <void_ptr>self._pvt_ptr[0] 24 ctx1jbcdefghiklmnopqrstuvwxy

137 def getPtr(self): (empty)

138 return <void_ptr>self._pvt_ptr 15 ctx1jklmnopqrstuvwx

139  

140cdef class anon_struct0: 

141 """ 

142 Attributes 

143 ---------- 

144  

145 available : int 

146  

147  

148  

149 compressedSize : size_t 

150  

151  

152  

153 uncompressedSize : size_t 

154  

155  

156  

157 cudaVersionMajor : int 

158  

159  

160  

161 cudaVersionMinor : int 

162  

163  

164  

165 numFiles : unsigned int 

166  

167  

168  

169 Methods 

170 ------- 

171 getPtr() 

172 Get memory address of class instance 

173 """ 

174 def __cinit__(self, void_ptr _ptr): 

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

176  

177 def __init__(self, void_ptr _ptr): 

178 pass 

179 def __dealloc__(self): 

180 pass 

181 def getPtr(self): (empty)

182 return <void_ptr>self._pvt_ptr 

183 def __repr__(self): 

184 if self._pvt_ptr is not NULL: 

185 str_list = [] 

186  

187 try: 

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

189 except ValueError: (empty)

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

191  

192  

193 try: 

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

195 except ValueError: 

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

197  

198  

199 try: 

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

201 except ValueError: 

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

203  

204  

205 try: 

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

207 except ValueError: 

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

209  

210  

211 try: 

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

213 except ValueError: 

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

215  

216  

217 try: 

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

219 except ValueError: 

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

221  

222 return '\n'.join(str_list) 

223 else: 

224 return '' 

225  

226 @property 

227 def available(self): 

228 return self._pvt_ptr[0].available (empty)

229 @available.setter 

230 def available(self, int available): 

231 self._pvt_ptr[0].available = available 

232  

233  

234 @property 

235 def compressedSize(self): 

236 return self._pvt_ptr[0].compressedSize 

237 @compressedSize.setter 

238 def compressedSize(self, size_t compressedSize): 

239 self._pvt_ptr[0].compressedSize = compressedSize 

240  

241  

242 @property 

243 def uncompressedSize(self): 

244 return self._pvt_ptr[0].uncompressedSize 

245 @uncompressedSize.setter 

246 def uncompressedSize(self, size_t uncompressedSize): 

247 self._pvt_ptr[0].uncompressedSize = uncompressedSize 

248  

249  

250 @property 

251 def cudaVersionMajor(self): 

252 return self._pvt_ptr[0].cudaVersionMajor 

253 @cudaVersionMajor.setter 

254 def cudaVersionMajor(self, int cudaVersionMajor): 

255 self._pvt_ptr[0].cudaVersionMajor = cudaVersionMajor 

256  

257  

258 @property 

259 def cudaVersionMinor(self): 

260 return self._pvt_ptr[0].cudaVersionMinor 

261 @cudaVersionMinor.setter 

262 def cudaVersionMinor(self, int cudaVersionMinor): 

263 self._pvt_ptr[0].cudaVersionMinor = cudaVersionMinor 

264  

265  

266 @property 

267 def numFiles(self): 

268 return self._pvt_ptr[0].numFiles 

269 @numFiles.setter 

270 def numFiles(self, unsigned int numFiles): 

271 self._pvt_ptr[0].numFiles = numFiles 

272  

273  

274cdef class nvrtcBundledHeadersInfo(anon_struct0): 

275 """ 

276 Attributes 

277 ---------- 

278  

279 available : int 

280  

281  

282  

283 compressedSize : size_t 

284  

285  

286  

287 uncompressedSize : size_t 

288  

289  

290  

291 cudaVersionMajor : int 

292  

293  

294  

295 cudaVersionMinor : int 

296  

297  

298  

299 numFiles : unsigned int 

300  

301  

302  

303 Methods 

304 ------- 

305 getPtr() 

306 Get memory address of class instance 

307 """ 

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

309 if _ptr == 0: 

310 self._pvt_ptr = <cynvrtc.nvrtcBundledHeadersInfo *>&self._pvt_val 

311 else: 

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

313  

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

315 pass 

316  

317@cython.embedsignature(True) (empty)

318def nvrtcGetErrorString(result not None : nvrtcResult): 

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

320  

321 Parameters 

322 ---------- 

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

324 CUDA Runtime Compilation API result code. 

325  

326 Returns 

327 ------- 

328 nvrtcResult.NVRTC_SUCCESS 

329 nvrtcResult.NVRTC_SUCCESS 

330 bytes 

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

332 """ 

333 cdef cynvrtc.nvrtcResult cyresult = int(result) 2 ctx2j .b

334 with nogil: 2 ctx2j .b

335 err = cynvrtc.nvrtcGetErrorString(cyresult) 2 ctx2j .b

336 return (nvrtcResult.NVRTC_SUCCESS, err) 2 ctx2j .b

337  

338@cython.embedsignature(True) (empty)

339def nvrtcVersion(): 

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

341  

342 Returns 

343 ------- 

344 nvrtcResult 

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

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

347 major : int 

348 CUDA Runtime Compilation major version number. 

349 minor : int 

350 CUDA Runtime Compilation minor version number. 

351 """ 

352 cdef int major = 0 143 ctx2a b c d e f g h i A B C D E 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

353 cdef int minor = 0 143 ctx2a b c d e f g h i A B C D E 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

354 with nogil: 143 ctx2a b c d e f g h i A B C D E 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

355 err = cynvrtc.nvrtcVersion(&major, &minor) 143 ctx2a b c d e f g h i A B C D E 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

356 if err != cynvrtc.NVRTC_SUCCESS: 143 ctx2a b c d e f g h i A B C D E 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

357 return (_nvrtcResult(err), None, None) 

358 return (_nvrtcResult_SUCCESS, major, minor) 143 ctx2a b c d e f g h i A B C D E 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

359  

360@cython.embedsignature(True) (empty)

361def nvrtcGetNumSupportedArchs(): 

362 """ 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. 

363  

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

365  

366 Returns 

367 ------- 

368 nvrtcResult 

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

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

371 numArchs : int 

372 number of supported architectures. 

373 """ 

374 cdef int numArchs = 0 1 ctx1z

375 with nogil: 1 ctx1z

376 err = cynvrtc.nvrtcGetNumSupportedArchs(&numArchs) 1 ctx1z

377 if err != cynvrtc.NVRTC_SUCCESS: 1 ctx1z

378 return (_nvrtcResult(err), None) 

379 return (_nvrtcResult_SUCCESS, numArchs) 1 ctx1z

380  

381@cython.embedsignature(True) (empty)

382def nvrtcGetSupportedArchs(): 

383 """ 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`. 

384  

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

386  

387 Returns 

388 ------- 

389 nvrtcResult 

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

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

392 supportedArchs : list[int] 

393 sorted array of supported architectures. 

394 """ 

395 cdef vector[int] supportedArchs 

396 _, s = nvrtcGetNumSupportedArchs() 1 ctx1z

397 supportedArchs.resize(s) 1 ctx1z

398  

399 with nogil: 1 ctx1z

400 err = cynvrtc.nvrtcGetSupportedArchs(supportedArchs.data()) 1 ctx1z

401 if err != cynvrtc.NVRTC_SUCCESS: 1 ctx1z

402 return (_nvrtcResult(err), None) 

403 return (_nvrtcResult_SUCCESS, supportedArchs) 1 ctx1z

404  

405@cython.embedsignature(True) (empty)

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

407 """ nvrtcCreateProgram creates an instance of nvrtcProgram with the given input parameters, and sets the output parameter `prog` with it. 

408  

409 Parameters 

410 ---------- 

411 src : bytes 

412 CUDA program source. 

413 name : bytes 

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

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

416 numHeaders : int 

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

418 to 0. 

419 headers : list[bytes] 

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

421 is 0. 

422 includeNames : list[bytes] 

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

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

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

426 here. 

427  

428 Returns 

429 ------- 

430 nvrtcResult 

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

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

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

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

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

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

437 CUDA Runtime Compilation program. 

438  

439 See Also 

440 -------- 

441 :py:obj:`~.nvrtcDestroyProgram` 

442 """ 

443 includeNames = [] if includeNames is None else includeNames 23 ctx1jbcdefghiklmnopqrstuvwx

444 if not all(isinstance(_x, (bytes)) for _x in includeNames): 23 ctx1jbcdefghiklmnopqrstuvwx

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

446 headers = [] if headers is None else headers 23 ctx1jbcdefghiklmnopqrstuvwx

447 if not all(isinstance(_x, (bytes)) for _x in headers): 23 ctx1jbcdefghiklmnopqrstuvwx

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

449 cdef nvrtcProgram prog = nvrtcProgram() 23 ctx1jbcdefghiklmnopqrstuvwx

450 if numHeaders > len(headers): raise RuntimeError("List is too small: " + str(len(headers)) + " < " + str(numHeaders)) 23 ctx1jbcdefghiklmnopqrstuvwx

451 if numHeaders > len(includeNames): raise RuntimeError("List is too small: " + str(len(includeNames)) + " < " + str(numHeaders)) 23 ctx1jbcdefghiklmnopqrstuvwx

452 cdef vector[const char*] cyheaders = headers 23 ctx1jbcdefghiklmnopqrstuvwx

453 cdef vector[const char*] cyincludeNames = includeNames 23 ctx1jbcdefghiklmnopqrstuvwx

454 with nogil: 23 ctx1jbcdefghiklmnopqrstuvwx

455 err = cynvrtc.nvrtcCreateProgram(<cynvrtc.nvrtcProgram*>prog._pvt_ptr, src, name, numHeaders, cyheaders.data(), cyincludeNames.data()) 23 ctx1jbcdefghiklmnopqrstuvwx

456 if err != cynvrtc.NVRTC_SUCCESS: 23 ctx1jbcdefghiklmnopqrstuvwx

457 return (_nvrtcResult(err), None) 

458 return (_nvrtcResult_SUCCESS, prog) 23 ctx1jbcdefghiklmnopqrstuvwx

459  

460@cython.embedsignature(True) (empty)

461def nvrtcDestroyProgram(prog): 

462 """ nvrtcDestroyProgram destroys the given program. 

463  

464 Parameters 

465 ---------- 

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

467 CUDA Runtime Compilation program. 

468  

469 Returns 

470 ------- 

471 nvrtcResult 

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

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

474  

475 See Also 

476 -------- 

477 :py:obj:`~.nvrtcCreateProgram` 

478 """ 

479 cdef cynvrtc.nvrtcProgram *cyprog 

480 if prog is None: 15 ctx1jklmnopqrstuvwx

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

482 elif isinstance(prog, (nvrtcProgram,)): 15 ctx1jklmnopqrstuvwx

483 pprog = prog.getPtr() 15 ctx1jklmnopqrstuvwx

484 cyprog = <cynvrtc.nvrtcProgram*><void_ptr>pprog 15 ctx1jklmnopqrstuvwx

485 elif isinstance(prog, (int)): 

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

487 else: 

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

489 with nogil: 15 ctx1jklmnopqrstuvwx

490 err = cynvrtc.nvrtcDestroyProgram(cyprog) 15 ctx1jklmnopqrstuvwx

491 return (_nvrtcResult(err),) 15 ctx1jklmnopqrstuvwx

492  

493@cython.embedsignature(True) (empty)

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

495 """ nvrtcCompileProgram compiles the given program. 

496  

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

498 Options`. 

499  

500 Parameters 

501 ---------- 

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

503 CUDA Runtime Compilation program. 

504 numOptions : int 

505 Number of compiler options passed. 

506 options : list[bytes] 

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

508 `NULL` when `numOptions` is 0. 

509  

510 Returns 

511 ------- 

512 nvrtcResult 

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

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

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

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

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

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

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

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

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

522 """ 

523 options = [] if options is None else options 23 ctx1jbcdefghiklmnopqrstuvwx

524 if not all(isinstance(_x, (bytes)) for _x in options): 23 ctx1jbcdefghiklmnopqrstuvwx

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

526 cdef cynvrtc.nvrtcProgram cyprog 

527 if prog is None: 23 ctx1jbcdefghiklmnopqrstuvwx

528 pprog = 0 

529 elif isinstance(prog, (nvrtcProgram,)): 23 ctx1jbcdefghiklmnopqrstuvwx

530 pprog = int(prog) 23 ctx1jbcdefghiklmnopqrstuvwx

531 else: 

532 pprog = int(nvrtcProgram(prog)) 

533 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 23 ctx1jbcdefghiklmnopqrstuvwx

534 if numOptions > len(options): raise RuntimeError("List is too small: " + str(len(options)) + " < " + str(numOptions)) 23 ctx1jbcdefghiklmnopqrstuvwx

535 cdef vector[const char*] cyoptions = options 23 ctx1jbcdefghiklmnopqrstuvwx

536 with nogil: 23 ctx1jbcdefghiklmnopqrstuvwx

537 err = cynvrtc.nvrtcCompileProgram(cyprog, numOptions, cyoptions.data()) 23 ctx1jbcdefghiklmnopqrstuvwx

538 return (_nvrtcResult(err),) 23 ctx1jbcdefghiklmnopqrstuvwx

539  

540@cython.embedsignature(True) (empty)

541def nvrtcGetPTXSize(prog): 

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

543  

544 Parameters 

545 ---------- 

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

547 CUDA Runtime Compilation program. 

548  

549 Returns 

550 ------- 

551 nvrtcResult 

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

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

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

555 ptxSizeRet : int 

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

557  

558 See Also 

559 -------- 

560 :py:obj:`~.nvrtcGetPTX` 

561 """ 

562 cdef cynvrtc.nvrtcProgram cyprog 

563 if prog is None: 

564 pprog = 0 

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

566 pprog = int(prog) 

567 else: 

568 pprog = int(nvrtcProgram(prog)) 

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

570 cdef size_t ptxSizeRet = 0 

571 with nogil: (empty)

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

573 if err != cynvrtc.NVRTC_SUCCESS: 

574 return (_nvrtcResult(err), None) 

575 return (_nvrtcResult_SUCCESS, ptxSizeRet) 

576  

577@cython.embedsignature(True) (empty)

578def nvrtcGetPTX(prog, char* ptx): 

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

580  

581 Parameters 

582 ---------- 

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

584 CUDA Runtime Compilation program. 

585 ptx : bytes 

586 Compiled result. 

587  

588 Returns 

589 ------- 

590 nvrtcResult 

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

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

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

594  

595 See Also 

596 -------- 

597 :py:obj:`~.nvrtcGetPTXSize` 

598 """ 

599 cdef cynvrtc.nvrtcProgram cyprog 

600 if prog is None: 

601 pprog = 0 

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

603 pprog = int(prog) 

604 else: 

605 pprog = int(nvrtcProgram(prog)) 

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

607 with nogil: 

608 err = cynvrtc.nvrtcGetPTX(cyprog, ptx) 

609 return (_nvrtcResult(err),) 

610  

611@cython.embedsignature(True) (empty)

612def nvrtcGetCUBINSize(prog): 

613 """ 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. 

614  

615 Parameters 

616 ---------- 

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

618 CUDA Runtime Compilation program. 

619  

620 Returns 

621 ------- 

622 nvrtcResult 

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

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

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

626 cubinSizeRet : int 

627 Size of the generated cubin. 

628  

629 See Also 

630 -------- 

631 :py:obj:`~.nvrtcGetCUBIN` 

632 """ 

633 cdef cynvrtc.nvrtcProgram cyprog 

634 if prog is None: 14 ctx1bcdefghiopqrst

635 pprog = 0 

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

637 pprog = int(prog) 14 ctx1bcdefghiopqrst

638 else: 

639 pprog = int(nvrtcProgram(prog)) 

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

641 cdef size_t cubinSizeRet = 0 14 ctx1bcdefghiopqrst

642 with nogil: 14 ctx1bcdefghiopqrst

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

644 if err != cynvrtc.NVRTC_SUCCESS: 14 ctx1bcdefghiopqrst

645 return (_nvrtcResult(err), None) 

646 return (_nvrtcResult_SUCCESS, cubinSizeRet) 14 ctx1bcdefghiopqrst

647  

648@cython.embedsignature(True) (empty)

649def nvrtcGetCUBIN(prog, char* cubin): 

650 """ 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. 

651  

652 Parameters 

653 ---------- 

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

655 CUDA Runtime Compilation program. 

656 cubin : bytes 

657 Compiled and assembled result. 

658  

659 Returns 

660 ------- 

661 nvrtcResult 

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

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

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

665  

666 See Also 

667 -------- 

668 :py:obj:`~.nvrtcGetCUBINSize` 

669 """ 

670 cdef cynvrtc.nvrtcProgram cyprog 

671 if prog is None: 14 ctx1bcdefghiopqrst

672 pprog = 0 

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

674 pprog = int(prog) 14 ctx1bcdefghiopqrst

675 else: 

676 pprog = int(nvrtcProgram(prog)) 

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

678 with nogil: 14 ctx1bcdefghiopqrst

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

680 return (_nvrtcResult(err),) 14 ctx1bcdefghiopqrst

681  

682@cython.embedsignature(True) (empty)

683def nvrtcGetLTOIRSize(prog): 

684 """ 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`. 

685  

686 Parameters 

687 ---------- 

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

689 CUDA Runtime Compilation program. 

690  

691 Returns 

692 ------- 

693 nvrtcResult 

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

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

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

697 LTOIRSizeRet : int 

698 Size of the generated LTO IR. 

699  

700 See Also 

701 -------- 

702 :py:obj:`~.nvrtcGetLTOIR` 

703 """ 

704 cdef cynvrtc.nvrtcProgram cyprog 

705 if prog is None: 8 ctx1klmnuvwx

706 pprog = 0 

707 elif isinstance(prog, (nvrtcProgram,)): 8 ctx1klmnuvwx

708 pprog = int(prog) 8 ctx1klmnuvwx

709 else: 

710 pprog = int(nvrtcProgram(prog)) 

711 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 8 ctx1klmnuvwx

712 cdef size_t LTOIRSizeRet = 0 8 ctx1klmnuvwx

713 with nogil: 8 ctx1klmnuvwx

714 err = cynvrtc.nvrtcGetLTOIRSize(cyprog, &LTOIRSizeRet) 8 ctx1klmnuvwx

715 if err != cynvrtc.NVRTC_SUCCESS: 8 ctx1klmnuvwx

716 return (_nvrtcResult(err), None) 

717 return (_nvrtcResult_SUCCESS, LTOIRSizeRet) 8 ctx1klmnuvwx

718  

719@cython.embedsignature(True) (empty)

720def nvrtcGetLTOIR(prog, char* LTOIR): 

721 """ 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`. 

722  

723 Parameters 

724 ---------- 

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

726 CUDA Runtime Compilation program. 

727 LTOIR : bytes 

728 Compiled result. 

729  

730 Returns 

731 ------- 

732 nvrtcResult 

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

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

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

736  

737 See Also 

738 -------- 

739 :py:obj:`~.nvrtcGetLTOIRSize` 

740 """ 

741 cdef cynvrtc.nvrtcProgram cyprog 

742 if prog is None: 8 ctx1klmnuvwx

743 pprog = 0 

744 elif isinstance(prog, (nvrtcProgram,)): 8 ctx1klmnuvwx

745 pprog = int(prog) 8 ctx1klmnuvwx

746 else: 

747 pprog = int(nvrtcProgram(prog)) 

748 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 8 ctx1klmnuvwx

749 with nogil: 8 ctx1klmnuvwx

750 err = cynvrtc.nvrtcGetLTOIR(cyprog, LTOIR) 8 ctx1klmnuvwx

751 return (_nvrtcResult(err),) 8 ctx1klmnuvwx

752  

753@cython.embedsignature(True) (empty)

754def nvrtcGetOptiXIRSize(prog): 

755 """ 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. 

756  

757 Parameters 

758 ---------- 

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

760 CUDA Runtime Compilation program. 

761  

762 Returns 

763 ------- 

764 nvrtcResult 

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

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

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

768 optixirSizeRet : int 

769 Size of the generated LTO IR. 

770  

771 See Also 

772 -------- 

773 :py:obj:`~.nvrtcGetOptiXIR` 

774 """ 

775 cdef cynvrtc.nvrtcProgram cyprog 

776 if prog is None: 

777 pprog = 0 

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

779 pprog = int(prog) 

780 else: 

781 pprog = int(nvrtcProgram(prog)) 

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

783 cdef size_t optixirSizeRet = 0 

784 with nogil: 

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

786 if err != cynvrtc.NVRTC_SUCCESS: 

787 return (_nvrtcResult(err), None) 

788 return (_nvrtcResult_SUCCESS, optixirSizeRet) 

789  

790@cython.embedsignature(True) (empty)

791def nvrtcGetOptiXIR(prog, char* optixir): 

792 """ 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. 

793  

794 Parameters 

795 ---------- 

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

797 CUDA Runtime Compilation program. 

798 optixir : bytes 

799 Optix IR Compiled result. 

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  

808 See Also 

809 -------- 

810 :py:obj:`~.nvrtcGetOptiXIRSize` 

811 """ 

812 cdef cynvrtc.nvrtcProgram cyprog 

813 if prog is None: 

814 pprog = 0 

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

816 pprog = int(prog) 

817 else: 

818 pprog = int(nvrtcProgram(prog)) 

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

820 with nogil: 

821 err = cynvrtc.nvrtcGetOptiXIR(cyprog, optixir) 

822 return (_nvrtcResult(err),) 

823  

824@cython.embedsignature(True) (empty)

825def nvrtcGetProgramLogSize(prog): 

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

827  

828 Note that compilation log may be generated with warnings and 

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

830  

831 Parameters 

832 ---------- 

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

834 CUDA Runtime Compilation program. 

835  

836 Returns 

837 ------- 

838 nvrtcResult 

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

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

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

842 logSizeRet : int 

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

844  

845 See Also 

846 -------- 

847 :py:obj:`~.nvrtcGetProgramLog` 

848 """ 

849 cdef cynvrtc.nvrtcProgram cyprog 

850 if prog is None: 9 ctx1jbcdefghi

851 pprog = 0 

852 elif isinstance(prog, (nvrtcProgram,)): 9 ctx1jbcdefghi

853 pprog = int(prog) 9 ctx1jbcdefghi

854 else: 

855 pprog = int(nvrtcProgram(prog)) 

856 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 9 ctx1jbcdefghi

857 cdef size_t logSizeRet = 0 9 ctx1jbcdefghi

858 with nogil: 9 ctx1jbcdefghi

859 err = cynvrtc.nvrtcGetProgramLogSize(cyprog, &logSizeRet) 9 ctx1jbcdefghi

860 if err != cynvrtc.NVRTC_SUCCESS: 9 ctx1jbcdefghi

861 return (_nvrtcResult(err), None) 

862 return (_nvrtcResult_SUCCESS, logSizeRet) 9 ctx1jbcdefghi

863  

864@cython.embedsignature(True) (empty)

865def nvrtcGetProgramLog(prog, char* log): 

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

867  

868 Parameters 

869 ---------- 

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

871 CUDA Runtime Compilation program. 

872 log : bytes 

873 Compilation log. 

874  

875 Returns 

876 ------- 

877 nvrtcResult 

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

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

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

881  

882 See Also 

883 -------- 

884 :py:obj:`~.nvrtcGetProgramLogSize` 

885 """ 

886 cdef cynvrtc.nvrtcProgram cyprog 

887 if prog is None: 9 ctx1jbcdefghi

888 pprog = 0 

889 elif isinstance(prog, (nvrtcProgram,)): 9 ctx1jbcdefghi

890 pprog = int(prog) 9 ctx1jbcdefghi

891 else: 

892 pprog = int(nvrtcProgram(prog)) 

893 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 9 ctx1jbcdefghi

894 with nogil: 9 ctx1jbcdefghi

895 err = cynvrtc.nvrtcGetProgramLog(cyprog, log) 9 ctx1jbcdefghi

896 return (_nvrtcResult(err),) 9 ctx1jbcdefghi

897  

898@cython.embedsignature(True) (empty)

899def nvrtcAddNameExpression(prog, char* name_expression): 

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

901  

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

903 call to nvrtcGetLoweredName to extract the lowered name. 

904  

905 Parameters 

906 ---------- 

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

908 CUDA Runtime Compilation program. 

909 name_expression : bytes 

910 constant expression denoting the address of a global function or 

911 device/__constant__ variable. 

912  

913 Returns 

914 ------- 

915 nvrtcResult 

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

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

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

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

920  

921 See Also 

922 -------- 

923 :py:obj:`~.nvrtcGetLoweredName` 

924 """ 

925 cdef cynvrtc.nvrtcProgram cyprog 

926 if prog is None: 

927 pprog = 0 

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

929 pprog = int(prog) 

930 else: 

931 pprog = int(nvrtcProgram(prog)) 

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

933 with nogil: 

934 err = cynvrtc.nvrtcAddNameExpression(cyprog, name_expression) 

935 return (_nvrtcResult(err),) 

936  

937@cython.embedsignature(True) (empty)

938def nvrtcGetLoweredName(prog, char* name_expression): 

939 """ 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. 

940  

941 Parameters 

942 ---------- 

943 prog : nvrtcProgram 

944 CUDA Runtime Compilation program. 

945 name_expression : bytes 

946 constant expression denoting the address of a global function or 

947 device/__constant__ variable. 

948  

949 Returns 

950 ------- 

951 nvrtcResult 

952 NVRTC_SUCCESS 

953 NVRTC_ERROR_NO_LOWERED_NAMES_BEFORE_COMPILATION 

954 NVRTC_ERROR_NAME_EXPRESSION_NOT_VALID 

955 lowered_name : bytes 

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

957 lowered (mangled) name corresponding to the provided name 

958 expression. 

959  

960 See Also 

961 -------- 

962 nvrtcAddNameExpression 

963 """ 

964 cdef cynvrtc.nvrtcProgram cyprog 

965 if prog is None: 1 ctx1y

966 pprog = 0 1 ctx1y

967 elif isinstance(prog, (nvrtcProgram,)): 1 ctx1y

968 pprog = int(prog) 

969 else: 

970 pprog = int(nvrtcProgram(prog)) 1 ctx1y

971 cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog 1 ctx1y

972 cdef const char* lowered_name = NULL 1 ctx1y

973 with nogil: 1 ctx1y

974 err = cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name) 1 ctx1y

975 if err != cynvrtc.NVRTC_SUCCESS: 1 ctx1y

976 return (_nvrtcResult(err), None) 1 ctx1y

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

978  

979@cython.embedsignature(True) (empty)

980def nvrtcGetPCHHeapSize(): 

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

982  

983 Returns 

984 ------- 

985 nvrtcResult 

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

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

988 ret : int 

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

990 """ 

991 cdef size_t ret = 0 

992 with nogil: 

993 err = cynvrtc.nvrtcGetPCHHeapSize(&ret) 

994 if err != cynvrtc.NVRTC_SUCCESS: 

995 return (_nvrtcResult(err), None) 

996 return (_nvrtcResult_SUCCESS, ret) 

997  

998@cython.embedsignature(True) (empty)

999def nvrtcSetPCHHeapSize(size_t size): 

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

1001  

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

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

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

1005  

1006 Parameters 

1007 ---------- 

1008 size : size_t 

1009 requested size of the PCH Heap, in bytes 

1010  

1011 Returns 

1012 ------- 

1013 nvrtcResult 

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

1015 """ 

1016 with nogil: 

1017 err = cynvrtc.nvrtcSetPCHHeapSize(size) 

1018 return (_nvrtcResult(err),) 

1019  

1020@cython.embedsignature(True) (empty)

1021def nvrtcGetPCHCreateStatus(prog): 

1022 """ returns the PCH creation status. 

1023  

1024 NVRTC_SUCCESS indicates that the PCH was successfully created. 

1025 NVRTC_ERROR_NO_PCH_CREATE_ATTEMPTED indicates that no PCH creation was 

1026 attempted, either because PCH functionality was not requested during 

1027 the preceding nvrtcCompileProgram call, or automatic PCH processing was 

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

1029 NVRTC_ERROR_PCH_CREATE_HEAP_EXHAUSTED indicates that a PCH file could 

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

1031 PCH heap. In this scenario, the 

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

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

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

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

1036 program instance. NVRTC_ERROR_PCH_CREATE indicates that an error 

1037 condition prevented the PCH file from being created. 

1038  

1039 Parameters 

1040 ---------- 

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

1042 CUDA Runtime Compilation program. 

1043  

1044 Returns 

1045 ------- 

1046 nvrtcResult 

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

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

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

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

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

1052 """ 

1053 cdef cynvrtc.nvrtcProgram cyprog 

1054 if prog is None: 

1055 pprog = 0 

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

1057 pprog = int(prog) 

1058 else: 

1059 pprog = int(nvrtcProgram(prog)) 

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

1061 with nogil: 

1062 err = cynvrtc.nvrtcGetPCHCreateStatus(cyprog) 

1063 return (_nvrtcResult(err),) 

1064  

1065@cython.embedsignature(True) (empty)

1066def nvrtcGetPCHHeapSizeRequired(prog): 

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

1068  

1069 Parameters 

1070 ---------- 

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

1072 CUDA Runtime Compilation program. 

1073  

1074 Returns 

1075 ------- 

1076 nvrtcResult 

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

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

1079 - :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 

1080 size : int 

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

1082 stored 

1083 """ 

1084 cdef cynvrtc.nvrtcProgram cyprog 

1085 if prog is None: 

1086 pprog = 0 

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

1088 pprog = int(prog) 

1089 else: 

1090 pprog = int(nvrtcProgram(prog)) 

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

1092 cdef size_t size = 0 

1093 with nogil: 

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

1095 if err != cynvrtc.NVRTC_SUCCESS: 

1096 return (_nvrtcResult(err), None) 

1097 return (_nvrtcResult_SUCCESS, size) 

1098  

1099@cython.embedsignature(True) (empty)

1100def nvrtcSetFlowCallback(prog, callback, payload): 

1101 """ 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. 

1102  

1103 The callback function must satisfy the following constraints: 

1104  

1105 (1) Its signature should be: 

1106  

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

1108  

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

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

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

1112 extensions. 

1113  

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

1115 return values are reserved for future use. 

1116  

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

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

1119 nvrtcCompileProgram call in progress. 

1120  

1121 (4) It must be thread-safe. 

1122  

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

1124  

1125 Parameters 

1126 ---------- 

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

1128 CUDA Runtime Compilation program. 

1129 callback : Any 

1130 the callback that issues cancellation signal. 

1131 payload : Any 

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

1133  

1134 Returns 

1135 ------- 

1136 nvrtcResult 

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

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

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

1140 """ 

1141 cdef cynvrtc.nvrtcProgram cyprog 

1142 if prog is None: 

1143 pprog = 0 

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

1145 pprog = int(prog) 

1146 else: 

1147 pprog = int(nvrtcProgram(prog)) 

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

1149 cdef _HelperInputVoidPtrStruct cycallbackHelper 

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

1151 cdef _HelperInputVoidPtrStruct cypayloadHelper 

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

1153 with nogil: 

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

1155 _helper_input_void_ptr_free(&cycallbackHelper) 

1156 _helper_input_void_ptr_free(&cypayloadHelper) 

1157 return (_nvrtcResult(err),) 

1158  

1159@cython.embedsignature(True) (empty)

1160def nvrtcGetTileIRSize(prog): 

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

1162  

1163 Parameters 

1164 ---------- 

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

1166 CUDA Runtime Compilation program. 

1167  

1168 Returns 

1169 ------- 

1170 nvrtcResult 

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

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

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

1174 TileIRSizeRet : int 

1175 Size of the generated cuda_tile IR. 

1176  

1177 See Also 

1178 -------- 

1179 :py:obj:`~.nvrtcGetTileIR` 

1180 """ 

1181 cdef cynvrtc.nvrtcProgram cyprog 

1182 if prog is None: 

1183 pprog = 0 

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

1185 pprog = int(prog) 

1186 else: 

1187 pprog = int(nvrtcProgram(prog)) 

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

1189 cdef size_t TileIRSizeRet = 0 

1190 with nogil: 

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

1192 if err != cynvrtc.NVRTC_SUCCESS: 

1193 return (_nvrtcResult(err), None) 

1194 return (_nvrtcResult_SUCCESS, TileIRSizeRet) 

1195  

1196@cython.embedsignature(True) (empty)

1197def nvrtcGetTileIR(prog, char* TileIR): 

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

1199  

1200 Parameters 

1201 ---------- 

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

1203 CUDA Runtime Compilation program. 

1204 TileIR : bytes 

1205 Generated cuda_tile IR. 

1206  

1207 Returns 

1208 ------- 

1209 nvrtcResult 

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

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

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

1213  

1214 See Also 

1215 -------- 

1216 :py:obj:`~.nvrtcGetTileIRSize` 

1217 """ 

1218 cdef cynvrtc.nvrtcProgram cyprog 

1219 if prog is None: 

1220 pprog = 0 

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

1222 pprog = int(prog) 

1223 else: 

1224 pprog = int(nvrtcProgram(prog)) 

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

1226 with nogil: 

1227 err = cynvrtc.nvrtcGetTileIR(cyprog, TileIR) 

1228 return (_nvrtcResult(err),) 

1229  

1230@cython.embedsignature(True) (empty)

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

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

1233  

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

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

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

1237 requiring a separate CUDA Toolkit installation. The bundled headers 

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

1239  

1240 After extraction, users can compile kernels by passing appropriate 

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

1242 nvrtcCompileProgram. 

1243  

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

1245 installation directory to track the installed version. 

1246  

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

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

1249 default, the function waits for the lock; use 

1250 NVRTC_INSTALL_HEADERS_NO_WAIT to return immediately with 

1251 NVRTC_ERROR_BUSY if another process holds the lock. 

1252  

1253 Parameters 

1254 ---------- 

1255 installPath : bytes 

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

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

1258 flags : unsigned int 

1259 NVRTC_INSTALL_HEADERS_* flags: 

1260  

1261 Returns 

1262 ------- 

1263 nvrtcResult 

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

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