diff options
Diffstat (limited to 'userspace/cuda/570.144')
| -rw-r--r-- | userspace/cuda/570.144/annotations/README.md | 3 | ||||
| -rw-r--r-- | userspace/cuda/570.144/annotations/name-candidates.tsv | 334 | ||||
| -rw-r--r-- | userspace/cuda/570.144/annotations/names.tsv | 39 | ||||
| -rw-r--r-- | userspace/cuda/570.144/annotations/subsystems.tsv | 49 | ||||
| -rw-r--r-- | userspace/cuda/570.144/ghidra/DecompileAt.java | 13 | ||||
| -rwxr-xr-x | userspace/cuda/570.144/scripts/ghidra_analyze.sh | 15 | ||||
| -rwxr-xr-x | userspace/cuda/570.144/scripts/ghidra_decompile_project.sh | 13 |
7 files changed, 460 insertions, 6 deletions
diff --git a/userspace/cuda/570.144/annotations/README.md b/userspace/cuda/570.144/annotations/README.md index 868c503..85c14af 100644 --- a/userspace/cuda/570.144/annotations/README.md +++ b/userspace/cuda/570.144/annotations/README.md @@ -8,8 +8,9 @@ accumulates across driver versions. | File | Format | Purpose | |---|---|---| -| `names.tsv` | `file_vaddr<TAB>name<TAB>source` | symbol map to apply to a Ghidra project (or any address-keyed analysis). `source` is `export`, `wrapper`, `verified`, or `todo`. | +| `names.tsv` | `file_vaddr<TAB>name<TAB>source` | symbol map to apply to a Ghidra project (or any address-keyed analysis). `source` is `export`, `wrapper`, `wrapper-callee`, `verified`, or `todo`. | | `subsystems.tsv` | `file_vaddr<TAB>name<TAB>subsystem<TAB>role<TAB>evidence` | human-readable subsystem index: which function belongs to which layer and what it does, with the evidence for the claim. | +| `name-candidates.tsv` | `file_vaddr<TAB>proposed_name<TAB>confidence<TAB>evidence<TAB>status` | machine-proposed API-implementation names for wrapper callees (`tools/name_candidates.py`), with the confidence and the candidate sets behind each proposal. `status` is `unverified` until the body is read; only `verified` entries may be promoted to `names.tsv`. | Addresses are **file virtual addresses** (`libcuda+0x...`). Apply the map with `tools/ghidra/ApplyNameMap.java`; see `docs/06-tooling.md`. diff --git a/userspace/cuda/570.144/annotations/name-candidates.tsv b/userspace/cuda/570.144/annotations/name-candidates.tsv new file mode 100644 index 0000000..559633f --- /dev/null +++ b/userspace/cuda/570.144/annotations/name-candidates.tsv @@ -0,0 +1,334 @@ +# Open-IOV CUDA API-implementation candidates (our own work product; tracked). +# Generated by tools/name_candidates.py from the elf-index exports, +# callgraph and the version's names.tsv; see that script for the method. +# +# file_vaddr<TAB>proposed_name<TAB>confidence<TAB>evidence<TAB>status +# confidence: high = exactly one filtered candidate, shared by >=2 variants +# or a unique callee; medium = one survivor also called by another +# wrapper, or several survivors (best-ranked proposed). +# status: unverified (tool output). Verification = read the body in the +# decompiled corpus and mark verified/corrected here; only verified +# names may be copied into names.tsv. +# +0x001a36d0 impl_cuCtxGetApiVersion high group=cuCtxGetApiVersion variants=cuCtxGetApiVersion; before=0x1a36d0,0x2bb140,0x2bb420,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1a36d0; outside_wrappers=0 verified +0x001a3770 impl_cuCtxGetId medium group=cuCtxGetId variants=cuCtxGetId; before=0x1a3770,0x2bb070,0x2bb140,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1a3770,0x2bb070; outside_wrappers=0 unverified +0x001a4480 impl_cuArrayGetSparseProperties medium group=cuArrayGetSparseProperties variants=cuArrayGetSparseProperties; before=0x1a4480,0x257720,0x258750,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a4480,0x258750,0x257720; outside_wrappers=0 unverified +0x001a46d0 impl_cuMipmappedArrayGetSparseProperties medium group=cuMipmappedArrayGetSparseProperties variants=cuMipmappedArrayGetSparseProperties; before=0x1a46d0,0x257720,0x258750,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a46d0,0x258750,0x257720; outside_wrappers=0 unverified +0x001a4900 impl_cuArrayGetMemoryRequirements high group=cuArrayGetMemoryRequirements variants=cuArrayGetMemoryRequirements; before=0x1a4900,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a4900; outside_wrappers=0 verified +0x001a4ac0 impl_cuMipmappedArrayGetMemoryRequirements high group=cuMipmappedArrayGetMemoryRequirements variants=cuMipmappedArrayGetMemoryRequirements; before=0x1a4ac0,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a4ac0; outside_wrappers=0 unverified +0x001a5410 impl_cuArrayGetPlane high group=cuArrayGetPlane variants=cuArrayGetPlane; before=0x1a5410,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a5410; outside_wrappers=0 verified +0x001a5630 impl_cuMipmappedArrayGetLevel high group=cuMipmappedArrayGetLevel variants=cuMipmappedArrayGetLevel; before=0x1a5630,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a5630; outside_wrappers=0 unverified +0x001a5820 impl_cuDevicePrimaryCtxGetState high group=cuDevicePrimaryCtxGetState variants=cuDevicePrimaryCtxGetState; before=0x1792f0,0x179350,0x1a5820,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1a5820; outside_wrappers=0 verified +0x001a6150 impl_cuDeviceRegisterAsyncNotification high group=cuDeviceRegisterAsyncNotification variants=cuDeviceRegisterAsyncNotification; before=0x1792f0,0x179350,0x179a80,0x1a6150,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a6150; outside_wrappers=0 verified +0x001a62c0 impl_cuDeviceUnregisterAsyncNotification high group=cuDeviceUnregisterAsyncNotification variants=cuDeviceUnregisterAsyncNotification; before=0x179200,0x1792f0,0x179350,0x1a62c0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a62c0; outside_wrappers=0 unverified +0x001a6430 impl_cuCtxAttach high group=cuCtxAttach variants=cuCtxAttach; before=0x1a6430,0x2bb140,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a6430; outside_wrappers=0 verified +0x001a66f0 impl_cuCtxGetCurrent high group=cuCtxGetCurrent variants=cuCtxGetCurrent; before=0x1798f0,0x1a66f0,0x2fde50,0x2fe440,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a66f0; outside_wrappers=0 verified +0x001a6850 impl_cuCtxGetDevice high group=cuCtxGetDevice variants=cuCtxGetDevice; before=0x1a6850,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a6850; outside_wrappers=0 verified +0x001a6960 impl_cuCtxGetFlags high group=cuCtxGetFlags variants=cuCtxGetFlags; before=0x1a6960,0x2dc100,0x2fde50,0x2fe540,0x4a7e90,0xd4bd20,0xd4dc50; after=0x1a6960; outside_wrappers=0 verified +0x001a6af0 impl_cuFlushGPUDirectRDMAWrites high group=cuFlushGPUDirectRDMAWrites variants=cuFlushGPUDirectRDMAWrites; before=0x1a6af0,0x2baed0,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a6af0; outside_wrappers=0 verified +0x001a6c50 impl_cuCtxGetExecAffinity high group=cuCtxGetExecAffinity variants=cuCtxGetExecAffinity; before=0x1a6c50,0x2baed0,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a6c50; outside_wrappers=0 verified +0x001a7320 impl_cuDeviceGetLuid high group=cuDeviceGetLuid variants=cuDeviceGetLuid; before=0x1a7320,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1a7320; outside_wrappers=0 verified +0x001a78c0 impl_cuDeviceGetDevResource high group=cuDeviceGetDevResource variants=cuDeviceGetDevResource; before=0x1a78c0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1a78c0; outside_wrappers=0 unverified +0x001a7aa0 impl_cuModuleGetFunctionCount high group=cuModuleGetFunctionCount variants=cuModuleGetFunctionCount; before=0x1a7aa0,0x2baed0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a7aa0; outside_wrappers=0 unverified +0x001a7c40 impl_cuModuleEnumerateFunctions high group=cuModuleEnumerateFunctions variants=cuModuleEnumerateFunctions; before=0x1a7c40,0x2baed0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a7c40; outside_wrappers=0 unverified +0x001a8120 impl_cuArrayDestroy medium group=cuArrayDestroy variants=cuArrayDestroy; before=0x1792f0,0x179350,0x1a8120,0x2597f0,0x2bb140,0x2d8f40,0x2e5900,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50,0xd4f3e0,0xd51980; after=0x1a8120,0x2597f0,0x2e5900; outside_wrappers=0 unverified +0x001a82f0 impl_cuMipmappedArrayCreate high group=cuMipmappedArrayCreate variants=cuMipmappedArrayCreate; before=0x1a82f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1a82f0; outside_wrappers=0 unverified +0x001a8520 impl_cuMipmappedArrayDestroy medium group=cuMipmappedArrayDestroy variants=cuMipmappedArrayDestroy; before=0x1792f0,0x179350,0x1a8520,0x25a070,0x2bb140,0x2d8f40,0x2e5a50,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50,0xd4f3e0,0xd51980; after=0x1a8520,0x2e5a50,0x25a070; outside_wrappers=0 unverified +0x001aa6c0 impl_cuCtxGetCacheConfig high group=cuCtxGetCacheConfig variants=cuCtxGetCacheConfig; before=0x179cf0,0x1aa6c0,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1aa6c0; outside_wrappers=0 unverified +0x001aa880 impl_cuCtxGetSharedMemConfig high group=cuCtxGetSharedMemConfig variants=cuCtxGetSharedMemConfig; before=0x179cf0,0x1aa880,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1aa880; outside_wrappers=0 unverified +0x001aad70 impl_cuCheckpointProcessLock medium group=cuCheckpointProcessLock variants=cuCheckpointProcessLock; before=0x179200,0x1aad70,0x2623e0,0x2fe540,0x30a870,0x4a7e90,0xd19d70,0xd4dc50; after=0x1aad70,0xd19d70,0x2623e0,0x30a870; outside_wrappers=0 unverified +0x001aae30 impl_cuCheckpointProcessCheckpoint high group=cuCheckpointProcessCheckpoint variants=cuCheckpointProcessCheckpoint; before=0x1aae30,0x2fe540,0x4a7e90,0xd4dc50; after=0x1aae30; outside_wrappers=0 unverified +0x001aaee0 impl_cuCheckpointProcessRestore high group=cuCheckpointProcessRestore variants=cuCheckpointProcessRestore; before=0x1aaee0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1aaee0; outside_wrappers=0 unverified +0x001aaf90 impl_cuCheckpointProcessUnlock high group=cuCheckpointProcessUnlock variants=cuCheckpointProcessUnlock; before=0x1aaf90,0x2fe540,0x4a7e90,0xd4dc50; after=0x1aaf90; outside_wrappers=0 unverified +0x001ab040 impl_cuCoredumpGetAttribute medium group=cuCoredumpGetAttribute variants=cuCoredumpGetAttribute; before=0x179190,0x179470,0x1ab040,0x2bb420,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ab040,0x179190,0x179470; outside_wrappers=0 unverified +0x001ab1e0 impl_cuCoredumpGetAttributeGlobal medium group=cuCoredumpGetAttributeGlobal variants=cuCoredumpGetAttributeGlobal; before=0x179190,0x179470,0x1ab1e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ab1e0,0x179190,0x179470; outside_wrappers=0 unverified +0x001ac3f0 impl_cuCtxDetach high group=cuCtxDetach variants=cuCtxDetach; before=0x1ac3f0,0x26d8e0,0x2bb300,0x2fde50,0x2fe0f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ac3f0; outside_wrappers=0 unverified +0x001acc50 impl_cuCtxSetCurrent medium group=cuCtxSetCurrent variants=cuCtxSetCurrent; before=0x1acc50,0x24a4d0,0x24b930,0x24b9e0,0x2fde50,0x2fde90,0x2fe540,0x4a7e90,0xd4dc50; after=0x1acc50,0x2fde90; outside_wrappers=0 unverified +0x001acec0 impl_cuCtxSetFlags medium group=cuCtxSetFlags variants=cuCtxSetFlags; before=0x1acec0,0x26ab80,0x2d8f40,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1acec0,0x26ab80; outside_wrappers=0 unverified +0x001ad4d0 impl_cuCtxResetPersistingL2Cache medium group=cuCtxResetPersistingL2Cache variants=cuCtxResetPersistingL2Cache; before=0x179cf0,0x1ad4d0,0x26b920,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1ad4d0,0x26b920; outside_wrappers=0 unverified +0x001ad670 impl_cuCtxSetCacheConfig medium group=cuCtxSetCacheConfig variants=cuCtxSetCacheConfig; before=0x1ad670,0x26ae40,0x2d8f40,0x2fe230,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1ad670,0x26ae40; outside_wrappers=0 unverified +0x001ad770 impl_cuCtxSetSharedMemConfig medium group=cuCtxSetSharedMemConfig variants=cuCtxSetSharedMemConfig; before=0x1ad770,0x26ae40,0x2d8f40,0x2fe230,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1ad770,0x26ae40; outside_wrappers=0 unverified +0x001ad870 impl_cuCtxEnablePeerAccess medium group=cuCtxEnablePeerAccess variants=cuCtxEnablePeerAccess; before=0x1792f0,0x179350,0x1ad870,0x2baed0,0x2c1500,0x2e5bf0,0x2fe230,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50; after=0x1ad870,0x2e5bf0,0x2c1500; outside_wrappers=0 unverified +0x001ada20 impl_cuCtxDisablePeerAccess medium group=cuCtxDisablePeerAccess variants=cuCtxDisablePeerAccess; before=0x1792f0,0x179350,0x1ada20,0x259530,0x2baed0,0x2d8f40,0x2e4ff0,0x2fe230,0x2fe540,0x443950,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1ada20,0x2e4ff0,0x259530,0x443950; outside_wrappers=0 unverified +0x001adc60 impl_cuCtxRecordEvent high group=cuCtxRecordEvent variants=cuCtxRecordEvent; before=0x1792f0,0x179350,0x1adc60,0x2886b0,0x2895f0,0x2bb140,0x2f8e10,0x2f8fd0,0x2fe540,0x4a7e90,0xd1e3f0,0xd4d9a0,0xd4dc50; after=0x1adc60; outside_wrappers=0 unverified +0x001ade30 impl_cuCtxWaitEvent medium group=cuCtxWaitEvent variants=cuCtxWaitEvent; before=0x1792f0,0x179350,0x1ade30,0x2bb140,0x2ed7d0,0x2fe540,0x4a7e90,0xd1e090,0xd1e230,0xd4d9a0,0xd4dc50; after=0x1ade30,0xd1e230; outside_wrappers=0 unverified +0x001adfb0 impl_cuDeviceGetByPCIBusId high group=cuDeviceGetByPCIBusId variants=cuDeviceGetByPCIBusId; before=0x1adfb0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1adfb0; outside_wrappers=0 unverified +0x001ae150 impl_cuDeviceGetPCIBusId medium group=cuDeviceGetPCIBusId variants=cuDeviceGetPCIBusId; before=0x179560,0x1ae150,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1ae150,0x179560; outside_wrappers=0 unverified +0x001ae290 impl_cuDeviceGetTexture1DLinearMaxWidth medium group=cuDeviceGetTexture1DLinearMaxWidth variants=cuDeviceGetTexture1DLinearMaxWidth; before=0x1ae290,0x257890,0x2583d0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1ae290,0x257890; outside_wrappers=0 unverified +0x001ae3d0 impl_cuDeviceComputeCapability high group=cuDeviceComputeCapability variants=cuDeviceComputeCapability; before=0x1ae3d0,0x26bc10,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1ae3d0; outside_wrappers=0 verified +0x001ae550 impl_cuDeviceGetProperties high group=cuDeviceGetProperties variants=cuDeviceGetProperties; before=0x1ae550,0x26bc10,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1ae550; outside_wrappers=0 verified +0x001ae780 impl_cuDeviceGetAttribute high group=cuDeviceGetAttribute variants=cuDeviceGetAttribute; before=0x1ae780,0x26bc10,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1ae780; outside_wrappers=0 verified +0x001aec20 impl_cuDeviceCanAccessPeer medium group=cuDeviceCanAccessPeer variants=cuDeviceCanAccessPeer; before=0x1aec20,0x2c1610,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1aec20,0x2c1610; outside_wrappers=0 unverified +0x001aed70 impl_cuDevicePrimaryCtxRetain medium group=cuDevicePrimaryCtxRetain variants=cuDevicePrimaryCtxRetain; before=0x1792f0,0x179350,0x1aed70,0x277570,0x2fe540,0x4a7e90,0xd4dc50; after=0x1aed70,0x277570; outside_wrappers=0 unverified +0x001af400 impl_cuDeviceSetMemPool medium group=cuDeviceSetMemPool variants=cuDeviceSetMemPool; before=0x1af400,0x2c3b00,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x1af400,0x2c3b00; outside_wrappers=0 unverified +0x001af5a0 impl_cuDeviceGetMemPool medium group=cuDeviceGetMemPool variants=cuDeviceGetMemPool; before=0x1af5a0,0x26bab0,0x2c61f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1af5a0,0x2c61f0,0x26bab0; outside_wrappers=0 unverified +0x001af960 impl_cuDeviceGetGraphMemAttribute high group=cuDeviceGetGraphMemAttribute variants=cuDeviceGetGraphMemAttribute; before=0x1af960,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1af960; outside_wrappers=0 unverified +0x001afae0 impl_cuDeviceSetGraphMemAttribute high group=cuDeviceSetGraphMemAttribute variants=cuDeviceSetGraphMemAttribute; before=0x1afae0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1afae0; outside_wrappers=0 unverified +0x001afc50 impl_cuDevSmResourceSplitByCount high group=cuDevSmResourceSplitByCount variants=cuDevSmResourceSplitByCount; before=0x1afc50,0x2fe540,0x4a7e90,0xd4dc50; after=0x1afc50; outside_wrappers=0 unverified +0x001afe80 impl_cuGraphicsEGLRegisterImage high group=cuGraphicsEGLRegisterImage variants=cuGraphicsEGLRegisterImage; before=0x1afe80,0x2fe540,0x4a7e90,0xd4dc50; after=0x1afe80; outside_wrappers=0 unverified +0x001b0130 impl_cuEGLStreamConsumerDisconnect medium group=cuEGLStreamConsumerDisconnect variants=cuEGLStreamConsumerDisconnect; before=0x1b0130,0x2fde50,0x2fe230,0x2fe540,0x350790,0x3c52d0,0x3c5370,0x400550,0x4a7e90,0xd4dc50; after=0x1b0130,0x350790; outside_wrappers=0 unverified +0x001b02b0 impl_cuEGLStreamProducerDisconnect medium group=cuEGLStreamProducerDisconnect variants=cuEGLStreamProducerDisconnect; before=0x1b02b0,0x2fde50,0x2fe230,0x2fe540,0x350790,0x3c52d0,0x3c5370,0x400550,0x4a7e90,0xd4dc50; after=0x1b02b0,0x350790; outside_wrappers=0 unverified +0x001b0440 impl_cuEGLStreamConsumerAcquireFrame high group=cuEGLStreamConsumerAcquireFrame variants=cuEGLStreamConsumerAcquireFrame; before=0x1b0440,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b0440; outside_wrappers=0 unverified +0x001b0960 impl_cuEGLStreamProducerConnect medium group=cuEGLStreamProducerConnect variants=cuEGLStreamProducerConnect; before=0x1b0960,0x26f650,0x270710,0x2fe230,0x2fe540,0x3c52d0,0x3c5370,0x3c54c0,0x4a7e90,0xd4da90,0xd4dc50; after=0x1b0960,0x270710,0x26f650,0x3c54c0; outside_wrappers=0 unverified +0x001b0af0 impl_cuEGLStreamProducerPresentFrame high group=cuEGLStreamProducerPresentFrame variants=cuEGLStreamProducerPresentFrame; before=0x1b0af0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b0af0; outside_wrappers=0 unverified +0x001b1370 impl_cuGraphDestroyNode high group=cuGraphDestroyNode variants=cuGraphDestroyNode; before=0x1b1370,0x2a2b90,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1b1370; outside_wrappers=0 unverified +0x001b1f20 impl_cuGraphAddEmptyNode high group=cuGraphAddEmptyNode variants=cuGraphAddEmptyNode; before=0x1b1f20,0x2a50f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b1f20; outside_wrappers=0 unverified +0x001b2070 impl_cuGraphAddEventRecordNode high group=cuGraphAddEventRecordNode variants=cuGraphAddEventRecordNode; before=0x1b2070,0x2a50f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b2070; outside_wrappers=0 unverified +0x001b2190 impl_cuGraphAddEventWaitNode high group=cuGraphAddEventWaitNode variants=cuGraphAddEventWaitNode; before=0x1b2190,0x2a50f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b2190; outside_wrappers=0 unverified +0x001b23c0 impl_cuEventQuery medium group=cuEventQuery variants=cuEventQuery; before=0x1b23c0,0x288ba0,0x2bb140,0x2fe230,0x2fe540,0x3034c0,0x4a7e90,0xd1e090,0xd4dc50; after=0x1b23c0,0x3034c0; outside_wrappers=0 unverified +0x001b3370 impl_cuIpcGetEventHandle medium group=cuIpcGetEventHandle variants=cuIpcGetEventHandle; before=0x1b3370,0x2bb420,0x2bbaa0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b3370,0x2bbaa0; outside_wrappers=0 unverified +0x001b37e0 impl_cuImportExternalMemory high group=cuImportExternalMemory variants=cuImportExternalMemory; before=0x1b37e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b37e0; outside_wrappers=0 unverified +0x001b3ab0 impl_cuExternalMemoryGetMappedBuffer medium group=cuExternalMemoryGetMappedBuffer variants=cuExternalMemoryGetMappedBuffer; before=0x1b3ab0,0x28a960,0x2d8f40,0x2fe540,0x4a7e90,0xd4d9a0,0xd4db70,0xd4dc50,0xd51980; after=0x1b3ab0,0x28a960; outside_wrappers=0 unverified +0x001b3cd0 impl_cuExternalMemoryGetMappedMipmappedArray high group=cuExternalMemoryGetMappedMipmappedArray variants=cuExternalMemoryGetMappedMipmappedArray; before=0x1b3cd0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b3cd0; outside_wrappers=0 unverified +0x001b3fb0 impl_cuDestroyExternalMemory medium group=cuDestroyExternalMemory variants=cuDestroyExternalMemory; before=0x1b3fb0,0x28b480,0x2d8f40,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1b3fb0,0x28b480; outside_wrappers=0 unverified +0x001b4110 impl_cuImportExternalSemaphore high group=cuImportExternalSemaphore variants=cuImportExternalSemaphore; before=0x1b4110,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b4110; outside_wrappers=0 unverified +0x001b50a0 impl_cuFuncSetBlockShape medium group=cuFuncSetBlockShape variants=cuFuncSetBlockShape; before=0x1b50a0,0x2938b0,0x2bb420,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1b50a0,0x2938b0; outside_wrappers=0 unverified +0x001b5260 impl_cuFuncSetSharedSize medium group=cuFuncSetSharedSize variants=cuFuncSetSharedSize; before=0x1b5260,0x293ae0,0x2bb420,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1b5260,0x293ae0; outside_wrappers=0 unverified +0x001b53f0 impl_cuFuncSetCacheConfig medium group=cuFuncSetCacheConfig variants=cuFuncSetCacheConfig; before=0x1b53f0,0x262ec0,0x263b80,0x263e10,0x264110,0x2baed0,0x2d8f40,0x2e0770,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1b53f0,0x262ec0,0x264110,0x263b80,0x263e10; outside_wrappers=0 unverified +0x001b5670 impl_cuFuncSetSharedMemConfig medium group=cuFuncSetSharedMemConfig variants=cuFuncSetSharedMemConfig; before=0x1b5670,0x262ec0,0x263b80,0x263e10,0x264110,0x2baed0,0x2d8f40,0x2e0770,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1b5670,0x262ec0,0x264110,0x263b80,0x263e10; outside_wrappers=0 unverified +0x001b58f0 impl_cuFuncGetModule high group=cuFuncGetModule variants=cuFuncGetModule; before=0x1b58f0,0x2baed0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b58f0; outside_wrappers=0 unverified +0x001b5a20 impl_cuFuncIsLoaded high group=cuFuncIsLoaded variants=cuFuncIsLoaded; before=0x1b5a20,0x2baed0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b5a20; outside_wrappers=0 unverified +0x001b5b70 impl_cuFuncLoad high group=cuFuncLoad variants=cuFuncLoad; before=0x1b5b70,0x2baed0,0x2e0770,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b5b70; outside_wrappers=0 unverified +0x001b5cb0 impl_cuFuncGetAttribute medium group=cuFuncGetAttribute variants=cuFuncGetAttribute; before=0x1b5cb0,0x293ec0,0x2baed0,0x2dc100,0x2fe540,0x4a7e90,0xd4bd20,0xd4dc50; after=0x1b5cb0,0x293ec0; outside_wrappers=0 unverified +0x001b5e80 impl_cuFuncSetAttribute medium group=cuFuncSetAttribute variants=cuFuncSetAttribute; before=0x1b5e80,0x293b70,0x2baed0,0x2d8f40,0x2e0770,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1b5e80,0x293b70; outside_wrappers=0 unverified +0x001b6030 impl_cuModuleGetFunction high group=cuModuleGetFunction variants=cuModuleGetFunction; before=0x1b6030,0x2baed0,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980,0xd6c4a0; after=0x1b6030; outside_wrappers=0 unverified +0x001b6210 impl_cuGraphicsGLRegisterBuffer medium group=cuGraphicsGLRegisterBuffer variants=cuGraphicsGLRegisterBuffer; before=0x179200,0x1799e0,0x1b6210,0x294fd0,0x2fe230,0x2fe540,0x3c52d0,0x3c5370,0x3c54c0,0x4a7e90,0xd4da90,0xd4dc50; after=0x1b6210,0x294fd0,0x3c54c0; outside_wrappers=0 unverified +0x001b63e0 impl_cuGraphicsGLRegisterImage high group=cuGraphicsGLRegisterImage variants=cuGraphicsGLRegisterImage; before=0x1b63e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b63e0; outside_wrappers=0 unverified +0x001b66e0 impl_cuGLRegisterBufferObject high group=cuGLRegisterBufferObject variants=cuGLRegisterBufferObject; before=0x1b66e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b66e0; outside_wrappers=0 unverified +0x001b75f0 impl_cuGraphNodeFindInClone high group=cuGraphNodeFindInClone variants=cuGraphNodeFindInClone; before=0x1b75f0,0x2fe540,0x3aa550,0x4a7e90,0xd4d9a0,0xd4dc50,0xd663e0; after=0x1b75f0; outside_wrappers=0 unverified +0x001b7e60 impl_cuGLUnregisterBufferObject high group=cuGLUnregisterBufferObject variants=cuGLUnregisterBufferObject; before=0x1b7e60,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b7e60; outside_wrappers=0 unverified +0x001b8090 impl_cuGLSetBufferObjectMapFlags high group=cuGLSetBufferObjectMapFlags variants=cuGLSetBufferObjectMapFlags; before=0x1b8090,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b8090; outside_wrappers=0 unverified +0x001b8580 impl_cuUserObjectRetain high group=cuUserObjectRetain variants=cuUserObjectRetain; before=0x179200,0x1b8580,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b8580; outside_wrappers=0 unverified +0x001b8670 impl_cuUserObjectRelease high group=cuUserObjectRelease variants=cuUserObjectRelease; before=0x179200,0x1b8670,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b8670; outside_wrappers=0 unverified +0x001b8760 impl_cuGraphRetainUserObject medium group=cuGraphRetainUserObject variants=cuGraphRetainUserObject; before=0x1792f0,0x179350,0x1b8760,0x2fe540,0x4a7e90,0xd4dc50,0xd6a780; after=0x1b8760,0xd6a780; outside_wrappers=0 unverified +0x001b8900 impl_cuGraphReleaseUserObject medium group=cuGraphReleaseUserObject variants=cuGraphReleaseUserObject; before=0x1792f0,0x179350,0x1b8900,0x2fe540,0x4a7e90,0xd4dc50,0xd6a780; after=0x1b8900,0xd6a780; outside_wrappers=0 unverified +0x001b8ba0 impl_cuGraphGetEdges medium group=cuGraphGetEdges variants=cuGraphGetEdges,cuGraphGetEdges_v2; before=0x1b8ba0,0x2a69b0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1b8ba0,0x2a69b0; outside_wrappers=0 unverified +0x001b8fd0 impl_cuGraphExecDestroy medium group=cuGraphExecDestroy variants=cuGraphExecDestroy; before=0x1b8fd0,0x1c1c60,0x2a3530,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b8fd0,0x2a3530; outside_wrappers=0 unverified +0x001b96b0 impl_cuGraphExecMemcpyNodeSetParams medium group=cuGraphExecMemcpyNodeSetParams variants=cuGraphExecMemcpyNodeSetParams; before=0x1b96b0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b96b0; outside_wrappers=1 unverified +0x001b9830 impl_cuGraphExecMemsetNodeSetParams medium group=cuGraphExecMemsetNodeSetParams variants=cuGraphExecMemsetNodeSetParams; before=0x1b9830,0x2fe540,0x4a7e90,0xd4dc50; after=0x1b9830; outside_wrappers=1 unverified +0x001ba6f0 impl_cuGraphAddHostNode high group=cuGraphAddHostNode variants=cuGraphAddHostNode; before=0x1ba6f0,0x2a63a0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ba6f0; outside_wrappers=0 unverified +0x001baec0 impl_cuGraphAddDependencies high group=cuGraphAddDependencies variants=cuGraphAddDependencies,cuGraphAddDependencies_v2; before=0x1baec0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1baec0; outside_wrappers=0 unverified +0x001bb0c0 impl_cuGraphRemoveDependencies high group=cuGraphRemoveDependencies variants=cuGraphRemoveDependencies,cuGraphRemoveDependencies_v2; before=0x1bb0c0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1bb0c0; outside_wrappers=0 unverified +0x001bb960 impl_cuMemcpyBatchAsync medium group=cuMemcpyBatchAsync variants=cuMemcpyBatchAsync_ptsz,cuMemcpyBatchAsync; before=0x1bb8a0,0x1bb960,0x1bddb0,0x2baed0,0x2f16c0,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50; after=0x1bb960,0x1bddb0,0x1bb8a0; outside_wrappers=0 unverified +0x001be0e0 impl_cuMemcpy3DBatchAsync high group=cuMemcpy3DBatchAsync variants=cuMemcpy3DBatchAsync_ptsz,cuMemcpy3DBatchAsync; before=0x1be0e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1be0e0; outside_wrappers=0 unverified +0x001be900 impl_cuCtxDestroy high group=cuCtxDestroy variants=cuCtxDestroy_v2,cuCtxDestroy; before=0x1be900,0x26d8e0,0x2baed0,0x2fde50,0x2fe0f0,0x2fe540,0x4a7e90,0xd4da50,0xd4dc50; after=0x1be900; outside_wrappers=0 verified +0x001bea40 impl_cuCtxPushCurrent medium group=cuCtxPushCurrent variants=cuCtxPushCurrent_v2,cuCtxPushCurrent; before=0x1bea40,0x24a4d0,0x24b930,0x2fde50,0x2fe020,0x2fe540,0x4a7e90,0xd4dc50,0xd4f3e0; after=0x1bea40,0x24b930; outside_wrappers=0 unverified +0x001bec50 impl_cuCtxPopCurrent medium group=cuCtxPopCurrent variants=cuCtxPopCurrent_v2,cuCtxPopCurrent; before=0x1bec50,0x24a4d0,0x24b9e0,0x2fde50,0x2fe0f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1bec50,0x24b9e0; outside_wrappers=0 unverified +0x001bee20 impl_cuDeviceGetNvSciSyncAttributes medium group=cuDeviceGetNvSciSyncAttributes variants=cuDeviceGetNvSciSyncAttributes; before=0x1bee20,0x1bf270,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1bee20,0x1bf270; outside_wrappers=0 unverified +0x001bf3e0 impl_cuDeviceGetDefaultMemPool medium group=cuDeviceGetDefaultMemPool variants=cuDeviceGetDefaultMemPool; before=0x1798f0,0x1bf3e0,0x1bf560,0x2fe440,0x2fe540,0x4a7e90,0xd4dc50; after=0x1bf3e0,0x1bf560; outside_wrappers=0 unverified +0x001bf660 impl_cuCtxGetDevResource high group=cuCtxGetDevResource variants=cuCtxGetDevResource; before=0x1bf660,0x2baed0,0x2bb140,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1bf660; outside_wrappers=0 unverified +0x001bf820 impl_cuGreenCtxGetDevResource high group=cuGreenCtxGetDevResource variants=cuGreenCtxGetDevResource; before=0x1bf820,0x2baed0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1bf820; outside_wrappers=0 unverified +0x001bf970 impl_cuDevResourceGenerateDesc high group=cuDevResourceGenerateDesc variants=cuDevResourceGenerateDesc; before=0x1bf970,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1bf970; outside_wrappers=0 unverified +0x001c0120 impl_cuEGLStreamConsumerConnect medium group=cuEGLStreamConsumerConnect variants=cuEGLStreamConsumerConnect; before=0x1c0120,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c0120; outside_wrappers=1 unverified +0x001c0120 impl_cuEGLStreamConsumerConnectWithFlags medium group=cuEGLStreamConsumerConnectWithFlags variants=cuEGLStreamConsumerConnectWithFlags; before=0x1c0120,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c0120; outside_wrappers=1 unverified +0x001c0930 impl_cuStreamBatchMemOp high group=cuStreamBatchMemOp variants=cuStreamBatchMemOp_v2_ptsz,cuStreamBatchMemOp,cuStreamBatchMemOp_ptsz,cuStreamBatchMemOp_v2; before=0x1c0930,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c0930; outside_wrappers=0 unverified +0x001c0d00 impl_cuGraphBatchMemOpNodeSetParams medium group=cuGraphBatchMemOpNodeSetParams variants=cuGraphBatchMemOpNodeSetParams; before=0x179200,0x179a80,0x179c00,0x1c0740,0x1c0d00,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c0d00; outside_wrappers=1 unverified +0x001c0f70 impl_cuSignalExternalSemaphoresAsync high group=cuSignalExternalSemaphoresAsync variants=cuSignalExternalSemaphoresAsync_ptsz,cuSignalExternalSemaphoresAsync; before=0x1c0f70,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c0f70; outside_wrappers=0 unverified +0x001c12c0 impl_cuWaitExternalSemaphoresAsync high group=cuWaitExternalSemaphoresAsync variants=cuWaitExternalSemaphoresAsync_ptsz,cuWaitExternalSemaphoresAsync; before=0x1c12c0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c12c0; outside_wrappers=0 unverified +0x001c15f0 impl_cuDestroyExternalSemaphore medium group=cuDestroyExternalSemaphore variants=cuDestroyExternalSemaphore; before=0x1798f0,0x1c15f0,0x1c1730,0x2fe230,0x2fe440,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c15f0,0x1c1730; outside_wrappers=0 unverified +0x001c1820 impl_cuGLCtxCreate high group=cuGLCtxCreate variants=cuGLCtxCreate_v2,cuGLCtxCreate; before=0x1c1820,0x2689e0,0x26d8e0,0x277680,0x2fe020,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c1820; outside_wrappers=0 unverified +0x001c2060 impl_cuGraphAddKernelNode high group=cuGraphAddKernelNode variants=cuGraphAddKernelNode_v2,cuGraphAddKernelNode; before=0x1c2060,0x1c2b10,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c2060; outside_wrappers=0 unverified +0x001c2570 impl_cuGraphAddMemsetNode medium group=cuGraphAddMemsetNode variants=cuGraphAddMemsetNode; before=0x1c2570,0x2a64e0,0x2baed0,0x2bb140,0x2ded70,0x2fde50,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c2570,0x2ded70,0x2a64e0; outside_wrappers=0 unverified +0x001c2830 impl_cuGraphMemsetNodeSetParams medium group=cuGraphMemsetNodeSetParams variants=cuGraphMemsetNodeSetParams; before=0x1c2830,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c2830; outside_wrappers=1 unverified +0x001c3c80 impl_cuGraphKernelNodeCopyAttributes high group=cuGraphKernelNodeCopyAttributes variants=cuGraphKernelNodeCopyAttributes; before=0x1c2b10,0x1c3c80,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c3c80; outside_wrappers=0 unverified +0x001c3e90 impl_cuGraphKernelNodeGetAttribute high group=cuGraphKernelNodeGetAttribute variants=cuGraphKernelNodeGetAttribute; before=0x1c3e90,0x2d8f40,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50,0xd51980; after=0x1c3e90; outside_wrappers=0 unverified +0x001c4540 impl_cuGraphKernelNodeSetAttribute medium group=cuGraphKernelNodeSetAttribute variants=cuGraphKernelNodeSetAttribute; before=0x1c40f0,0x1c4540,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1c4540,0x1c40f0; outside_wrappers=0 unverified +0x001c4a90 impl_cuGraphExecNodeSetParams medium group=cuGraphExecNodeSetParams variants=cuGraphExecNodeSetParams; before=0x1a39d0,0x1b96b0,0x1b9830,0x1b9a90,0x1b9ea0,0x1c0e20,0x1c3870,0x1c4920,0x1c4a90,0x2ac1b0,0x2b4f80,0x2fe540,0x4a7e90,0xd183b0,0xd4dc50; after=0x1c4a90,0x1b9a90,0x1b9ea0,0x1a39d0,0x2b4f80,0x1c3870,0x1b96b0,0x1c0e20,0x1b9830,0x2ac1b0,0x1c4920; outside_wrappers=0 unverified +0x001c4d50 impl_cuGraphMemcpyNodeSetParams medium group=cuGraphMemcpyNodeSetParams variants=cuGraphMemcpyNodeSetParams; before=0x1c4d50,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c4d50; outside_wrappers=1 unverified +0x001c4e80 impl_cuGraphNodeSetParams medium group=cuGraphNodeSetParams variants=cuGraphNodeSetParams; before=0x1c0d00,0x1c2830,0x1c2a00,0x1c3750,0x1c4920,0x1c4d50,0x1c4e80,0x2a9f60,0x2aa190,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c4e80,0x1c2a00,0x1c2830,0x1c0d00,0x1c4d50,0x1c3750,0x1c4920; outside_wrappers=0 unverified +0x001c51c0 impl_cuGraphAddMemcpyNode medium group=cuGraphAddMemcpyNode variants=cuGraphAddMemcpyNode; before=0x1c51c0,0x2a60a0,0x2baed0,0x2bb140,0x2d7780,0x2fde50,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c51c0,0x2a60a0,0x2d7780; outside_wrappers=0 unverified +0x001c53e0 impl_cuGraphAddNode high group=cuGraphAddNode variants=cuGraphAddNode,cuGraphAddNode_v2; before=0x1c53e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c53e0; outside_wrappers=0 unverified +0x001c6360 impl_cuMemExportToShareableHandle high group=cuMemExportToShareableHandle variants=cuMemExportToShareableHandle; before=0x1c6360,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c6360; outside_wrappers=0 unverified +0x001c64f0 impl_cuLibraryGetKernel high group=cuLibraryGetKernel variants=cuLibraryGetKernel; before=0x1c64f0,0x2fe540,0x3aa550,0x4a7e90,0xd4d9a0,0xd4dc50,0xd663e0; after=0x1c64f0; outside_wrappers=0 unverified +0x001c6610 impl_cuLibraryGetManaged high group=cuLibraryGetManaged variants=cuLibraryGetManaged; before=0x1c6610,0x2fe540,0x3aa550,0x4a7e90,0xd4d9a0,0xd4dc50,0xd663e0; after=0x1c6610; outside_wrappers=0 unverified +0x001c69f0 impl_cuMemGetAllocationGranularity high group=cuMemGetAllocationGranularity variants=cuMemGetAllocationGranularity; before=0x1c69f0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c69f0; outside_wrappers=0 unverified +0x001c6ec0 impl_cuGreenCtxRecordEvent high group=cuGreenCtxRecordEvent variants=cuGreenCtxRecordEvent; before=0x1792f0,0x179350,0x1c6ec0,0x2886b0,0x2895f0,0x2baed0,0x2f8e10,0x2f8fd0,0x2fe540,0x4a7e90,0xd1e3f0,0xd4d9a0,0xd4dc50; after=0x1c6ec0; outside_wrappers=0 unverified +0x001c7000 impl_cuGreenCtxWaitEvent medium group=cuGreenCtxWaitEvent variants=cuGreenCtxWaitEvent; before=0x1c7000,0x2baed0,0x2ed7d0,0x2fe540,0x4a7e90,0xd1e230,0xd1e3f0,0xd4d9a0,0xd4dc50; after=0x1c7000,0xd1e230; outside_wrappers=0 unverified +0x001c7120 impl_cuGreenCtxStreamCreate medium group=cuGreenCtxStreamCreate variants=cuGreenCtxStreamCreate; before=0x1c7120,0x2baed0,0x2ef570,0x2f0fb0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50,0xd613b0; after=0x1c7120,0x2ef570,0x2f0fb0,0xd613b0; outside_wrappers=0 unverified +0x001c73e0 impl_cuGraphicsResourceGetMappedMipmappedArray high group=cuGraphicsResourceGetMappedMipmappedArray variants=cuGraphicsResourceGetMappedMipmappedArray; before=0x1798f0,0x1c73e0,0x2baed0,0x2fe440,0x2fe540,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50; after=0x1c73e0; outside_wrappers=0 unverified +0x001c7530 impl_cuGraphicsSubResourceGetMappedArray high group=cuGraphicsSubResourceGetMappedArray variants=cuGraphicsSubResourceGetMappedArray; before=0x1c7530,0x2baed0,0x2fe540,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50; after=0x1c7530; outside_wrappers=0 unverified +0x001c7cf0 impl_cuMemPoolGetAccess high group=cuMemPoolGetAccess variants=cuMemPoolGetAccess; before=0x1c7cf0,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c7cf0; outside_wrappers=0 unverified +0x001c7ed0 impl_cuGraphicsUnregisterResource high group=cuGraphicsUnregisterResource variants=cuGraphicsUnregisterResource; before=0x1c7ed0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1c7ed0; outside_wrappers=0 unverified +0x001c80b0 impl_cuGraphicsResourceGetMappedEglFrame high group=cuGraphicsResourceGetMappedEglFrame variants=cuGraphicsResourceGetMappedEglFrame; before=0x1c80b0,0x2baed0,0x2fde50,0x2fe540,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50; after=0x1c80b0; outside_wrappers=0 unverified +0x001c94b0 impl_cuMemPoolExportPointer high group=cuMemPoolExportPointer variants=cuMemPoolExportPointer; before=0x1792f0,0x179350,0x1c94b0,0x2fe230,0x2fe540,0x3fe100,0x4a7e90,0xd4dc50; after=0x1c94b0; outside_wrappers=0 unverified +0x001c9700 impl_cuMemGetAllocationPropertiesFromHandle high group=cuMemGetAllocationPropertiesFromHandle variants=cuMemGetAllocationPropertiesFromHandle; before=0x1c9700,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c9700; outside_wrappers=0 unverified +0x001c9b90 impl_cuKernelGetParamInfo high group=cuKernelGetParamInfo variants=cuKernelGetParamInfo; before=0x1c9b90,0x2bb420,0x2bfad0,0x2fe540,0x4a7e90,0xd4dc50,0xd6c4a0; after=0x1c9b90; outside_wrappers=0 unverified +0x001c9d80 impl_cuLibraryLoadFromFile medium group=cuLibraryLoadFromFile variants=cuLibraryLoadFromFile; before=0x179200,0x1c9d80,0x2bbd30,0x2d3f60,0x2df390,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1c9d80,0x2d3f60; outside_wrappers=0 unverified +0x001c9f90 impl_cuLibraryGetGlobal high group=cuLibraryGetGlobal variants=cuLibraryGetGlobal; before=0x1c9f90,0x2bb420,0x2bfad0,0x2f9f00,0x2fe540,0x402500,0x4a7e90,0xd4dc50; after=0x1c9f90; outside_wrappers=0 unverified +0x001ca790 impl_cuCtxGetLimit medium group=cuCtxGetLimit variants=cuCtxGetLimit; before=0x1ca790,0x2bb420,0x2c0400,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1ca790,0x2c0400; outside_wrappers=0 unverified +0x001ca940 impl_cuMemHostUnregister high group=cuMemHostUnregister variants=cuMemHostUnregister; before=0x1ca940,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ca940; outside_wrappers=0 unverified +0x001cab60 impl_cuMemAllocManaged high group=cuMemAllocManaged variants=cuMemAllocManaged; before=0x1cab60,0x2fe540,0x4a7e90,0xd4dc50; after=0x1cab60; outside_wrappers=0 unverified +0x001cb030 impl_cuMemRangeGetAttributes medium group=cuMemRangeGetAttributes variants=cuMemRangeGetAttributes; before=0x1cb030,0x2fe540,0x4a7e90,0xd4dc50; after=0x1cb030; outside_wrappers=1 unverified +0x001cbda0 impl_cuMemHostGetFlags high group=cuMemHostGetFlags variants=cuMemHostGetFlags; before=0x1cbda0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1cbda0; outside_wrappers=0 unverified +0x001cc6f0 impl_cuIpcGetMemHandle medium group=cuIpcGetMemHandle variants=cuIpcGetMemHandle; before=0x1cc6f0,0x2bb420,0x2bbaa0,0x2fe540,0x3fe700,0x402500,0x4a7e90,0xd4bd90,0xd4dc50; after=0x1cc6f0,0x2bbaa0; outside_wrappers=0 unverified +0x001ccb30 impl_cuIpcCloseMemHandle high group=cuIpcCloseMemHandle variants=cuIpcCloseMemHandle; before=0x1ccb30,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ccb30; outside_wrappers=0 unverified +0x001ccd30 impl_cuMemGetHandleForAddressRange high group=cuMemGetHandleForAddressRange variants=cuMemGetHandleForAddressRange; before=0x1795e0,0x1ccd30,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ccd30; outside_wrappers=0 unverified +0x001d2a60 impl_cuMemAddressReserve medium group=cuMemAddressReserve variants=cuMemAddressReserve; before=0x1792f0,0x179350,0x1795e0,0x1d2a60,0x2d8ae0,0x2fe540,0x3acc80,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1d2a60,0x3acc80,0x2d8ae0; outside_wrappers=0 unverified +0x001d2c30 impl_cuMemAddressFree medium group=cuMemAddressFree variants=cuMemAddressFree; before=0x1792f0,0x179350,0x1795e0,0x1d2c30,0x2d8c30,0x2fe540,0x306160,0x3ac0f0,0x3ac3f0,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1d2c30,0x2d8c30,0x3ac0f0; outside_wrappers=0 unverified +0x001d2f80 impl_cuMemPoolImportPointer high group=cuMemPoolImportPointer variants=cuMemPoolImportPointer; before=0x1d2f80,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d2f80; outside_wrappers=0 unverified +0x001d3bf0 impl_cuMemUnmap high group=cuMemUnmap variants=cuMemUnmap; before=0x1d3bf0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d3bf0; outside_wrappers=0 unverified +0x001d4010 impl_cuMemImportFromShareableHandle high group=cuMemImportFromShareableHandle variants=cuMemImportFromShareableHandle; before=0x1d4010,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d4010; outside_wrappers=0 unverified +0x001d4800 impl_cuMemRetainAllocationHandle medium group=cuMemRetainAllocationHandle variants=cuMemRetainAllocationHandle; before=0x1792f0,0x179350,0x1d4800,0x257620,0x2fe540,0x3ac3f0,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1d4800,0x257620; outside_wrappers=0 unverified +0x001d5020 impl_cuMemPoolTrimTo medium group=cuMemPoolTrimTo variants=cuMemPoolTrimTo; before=0x1792f0,0x179350,0x1d5020,0x2c51a0,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d5020,0x2c51a0; outside_wrappers=0 unverified +0x001d5180 impl_cuMemPoolSetAttribute medium group=cuMemPoolSetAttribute variants=cuMemPoolSetAttribute; before=0x1792f0,0x179350,0x1d5180,0x2db020,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d5180,0x2db020; outside_wrappers=0 unverified +0x001d52d0 impl_cuMemPoolGetAttribute medium group=cuMemPoolGetAttribute variants=cuMemPoolGetAttribute; before=0x1792f0,0x179350,0x1d52d0,0x2db110,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d52d0,0x2db110; outside_wrappers=0 unverified +0x001d5430 impl_cuMemPoolSetAccess high group=cuMemPoolSetAccess variants=cuMemPoolSetAccess; before=0x1d5430,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d5430; outside_wrappers=0 unverified +0x001d5a00 impl_cuMemPoolExportToShareableHandle medium group=cuMemPoolExportToShareableHandle variants=cuMemPoolExportToShareableHandle; before=0x1792f0,0x179350,0x1799e0,0x1d5a00,0x2c6e60,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d5a00,0x2c6e60; outside_wrappers=0 unverified +0x001d5c60 impl_cuMemPoolImportFromShareableHandle medium group=cuMemPoolImportFromShareableHandle variants=cuMemPoolImportFromShareableHandle; before=0x1792f0,0x179350,0x179450,0x1799e0,0x1d5c60,0x28d350,0x2c3b00,0x2fe230,0x2fe540,0x30a9d0,0x3aa550,0x3ab420,0x4a7e90,0xd4dc50,0xd663e0; after=0x1d5c60,0x28d350,0x3ab420,0x30a9d0,0x2c3b00; outside_wrappers=0 unverified +0x001d5ed0 impl_cuGraphExecUpdate high group=cuGraphExecUpdate variants=cuGraphExecUpdate_v2,cuGraphExecUpdate; before=0x1d5ed0,0x2fe540,0x4a7e90,0xd4da50,0xd4dc50; after=0x1d5ed0; outside_wrappers=0 unverified +0x001d61c0 impl_cuGraphConditionalHandleCreate high group=cuGraphConditionalHandleCreate variants=cuGraphConditionalHandleCreate; before=0x1d61c0,0x2baed0,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d61c0; outside_wrappers=0 unverified +0x001d6410 impl_cuGraphicsResourceSetMapFlags high group=cuGraphicsResourceSetMapFlags variants=cuGraphicsResourceSetMapFlags_v2,cuGraphicsResourceSetMapFlags; before=0x1d6410,0x2baed0,0x2fe540,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50; after=0x1d6410; outside_wrappers=0 unverified +0x001d65a0 impl_cuGraphicsMapResources high group=cuGraphicsMapResources variants=cuGraphicsMapResources_ptsz,cuGraphicsMapResources; before=0x1d65a0,0x2baed0,0x2f16c0,0x2fe230,0x2fe540,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50; after=0x1d65a0; outside_wrappers=0 unverified +0x001d67a0 impl_cuGraphicsUnmapResources high group=cuGraphicsUnmapResources variants=cuGraphicsUnmapResources_ptsz,cuGraphicsUnmapResources; before=0x1d67a0,0x2baed0,0x2f16c0,0x2fe230,0x2fe540,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50; after=0x1d67a0; outside_wrappers=0 unverified +0x001d69a0 impl_cuLinkAddFile medium group=cuLinkAddFile variants=cuLinkAddFile_v2,cuLinkAddFile; before=0x179200,0x1d69a0,0x2bbd30,0x2d2990,0x2df390,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d69a0,0x2df390; outside_wrappers=0 unverified +0x001d7d70 impl_cuLaunchCooperativeKernelMultiDevice high group=cuLaunchCooperativeKernelMultiDevice variants=cuLaunchCooperativeKernelMultiDevice; before=0x1d7d70,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d7d70; outside_wrappers=0 unverified +0x001d9540 impl_cuKernelSetCacheConfig high group=cuKernelSetCacheConfig variants=cuKernelSetCacheConfig; before=0x1d9540,0x2bb420,0x2fe540,0x4a7e90,0xd4dc50; after=0x1d9540; outside_wrappers=0 unverified +0x001da450 impl_cuMemFree high group=cuMemFree variants=cuMemFree_v2,cuMemFree; before=0x1da450,0x2fe540,0x4a7e90,0xd4dc50; after=0x1da450; outside_wrappers=0 verified +0x001da880 impl_cuMemFreeHost high group=cuMemFreeHost variants=cuMemFreeHost; before=0x1da880,0x2fe540,0x4a7e90,0xd4dc50; after=0x1da880; outside_wrappers=0 verified +0x001daad0 impl_cuMemAllocHost high group=cuMemAllocHost variants=cuMemAllocHost_v2,cuMemAllocHost; before=0x1daad0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1daad0; outside_wrappers=1 unverified +0x001dae90 impl_cuMemPrefetchAsync high group=cuMemPrefetchAsync variants=cuMemPrefetchAsync_ptsz,cuMemPrefetchAsync_v2_ptsz,cuMemPrefetchAsync,cuMemPrefetchAsync_v2; before=0x1dae90,0x2fe540,0x4a7e90,0xd4dc50; after=0x1dae90; outside_wrappers=0 unverified +0x001db0f0 impl_cuMemAdvise high group=cuMemAdvise variants=cuMemAdvise,cuMemAdvise_v2; before=0x1db0f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1db0f0; outside_wrappers=0 unverified +0x001dbdd0 impl_cuPointerGetAttributes medium group=cuPointerGetAttributes variants=cuPointerGetAttributes; before=0x1db310,0x1dbdd0,0x2bb300,0x2fe540,0x3fe100,0x402500,0x4a7e90,0xd4dc50; after=0x1dbdd0,0x1db310; outside_wrappers=0 unverified +0x001dc060 impl_cuMemHostGetDevicePointer medium group=cuMemHostGetDevicePointer variants=cuMemHostGetDevicePointer_v2,cuMemHostGetDevicePointer; before=0x1792f0,0x179350,0x1dc060,0x2bb300,0x2fe540,0x3ac3f0,0x4026d0,0x403470,0x4a7e90,0xd4dc50; after=0x1dc060,0x4026d0; outside_wrappers=0 unverified +0x001dc310 impl_cuMemHostRegister high group=cuMemHostRegister variants=cuMemHostRegister_v2,cuMemHostRegister; before=0x1dc310,0x2fe540,0x4a7e90,0xd4dc50; after=0x1dc310; outside_wrappers=0 unverified +0x001dc640 impl_cuIpcOpenMemHandle medium group=cuIpcOpenMemHandle variants=cuIpcOpenMemHandle_v2,cuIpcOpenMemHandle; before=0x1dc640,0x2bb420,0x2c1d40,0x2d8f40,0x2e7bf0,0x2fe540,0x400550,0x403330,0x4a7e90,0xd4dc50,0xd51980; after=0x1dc640,0x403330; outside_wrappers=0 unverified +0x001dc7c0 impl_cuPointerSetAttribute medium group=cuPointerSetAttribute variants=cuPointerSetAttribute; before=0x1dc7c0,0x1dcc50,0x2fde50,0x2fe230,0x2fe540,0x3fe100,0x402500,0x4a7e90,0xd4dc50; after=0x1dc7c0,0x1dcc50; outside_wrappers=0 unverified +0x001dc9b0 impl_cuPointerGetAttribute high group=cuPointerGetAttribute variants=cuPointerGetAttribute; before=0x1dc9b0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1dc9b0; outside_wrappers=0 unverified +0x001dfbe0 impl_cuMemSetAccess medium group=cuMemSetAccess variants=cuMemSetAccess; before=0x1dfbe0,0x1e0420,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1dfbe0,0x1e0420; outside_wrappers=0 unverified +0x001e0640 impl_cuMemGetAccess medium group=cuMemGetAccess variants=cuMemGetAccess; before=0x1e0640,0x1e0950,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1e0640,0x1e0950; outside_wrappers=0 unverified +0x001e0b40 impl_cuMemAllocFromPoolAsync high group=cuMemAllocFromPoolAsync variants=cuMemAllocFromPoolAsync_ptsz,cuMemAllocFromPoolAsync; before=0x1e0b40,0x2f16c0,0x2fe230,0x2fe540,0x3c7f40,0x4a7e90,0xd4dc50; after=0x1e0b40; outside_wrappers=0 unverified +0x001e0cb0 impl_cuMemCreate high group=cuMemCreate variants=cuMemCreate; before=0x1e0cb0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1e0cb0; outside_wrappers=0 unverified +0x001e18f0 impl_cuMemMap high group=cuMemMap variants=cuMemMap; before=0x1e18f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1e18f0; outside_wrappers=0 unverified +0x001e9ab0 impl_cuModuleLoad high group=cuModuleLoad variants=cuModuleLoad; before=0x1e9ab0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1e9ab0; outside_wrappers=0 verified +0x001e9d80 impl_cuModuleLoadData high group=cuModuleLoadData variants=cuModuleLoadData; before=0x1e9d80,0x2fe540,0x4a7e90,0xd4dc50; after=0x1e9d80; outside_wrappers=0 verified +0x001ea020 impl_cuModuleLoadFatBinary high group=cuModuleLoadFatBinary variants=cuModuleLoadFatBinary; before=0x1ea020,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ea020; outside_wrappers=0 unverified +0x001ea370 impl_cuTensorMapEncodeTiled high group=cuTensorMapEncodeTiled variants=cuTensorMapEncodeTiled; before=0x1ea370,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50; after=0x1ea370; outside_wrappers=0 unverified +0x001ea530 impl_cuTensorMapEncodeIm2col high group=cuTensorMapEncodeIm2col variants=cuTensorMapEncodeIm2col; before=0x1ea530,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50; after=0x1ea530; outside_wrappers=0 unverified +0x001ea6f0 impl_cuTensorMapEncodeIm2colWide high group=cuTensorMapEncodeIm2colWide variants=cuTensorMapEncodeIm2colWide; before=0x1ea6f0,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50; after=0x1ea6f0; outside_wrappers=0 unverified +0x001ea8f0 impl_cuTensorMapReplaceAddress high group=cuTensorMapReplaceAddress variants=cuTensorMapReplaceAddress; before=0x1798f0,0x1ea8f0,0x2fe440,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50; after=0x1ea8f0; outside_wrappers=0 unverified +0x001ea9d0 impl_cuModuleLoadDataEx high group=cuModuleLoadDataEx variants=cuModuleLoadDataEx; before=0x1ea9d0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ea9d0; outside_wrappers=0 unverified +0x001eb4d0 impl_cuSurfRefGetArray high group=cuSurfRefGetArray variants=cuSurfRefGetArray; before=0x1eb4d0,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x1eb4d0; outside_wrappers=0 unverified +0x001eb960 impl_cuModuleGetTexRef medium group=cuModuleGetTexRef variants=cuModuleGetTexRef; before=0x179980,0x1eb960,0x2baed0,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1eb960,0x179980; outside_wrappers=0 unverified +0x001ebb60 impl_cuModuleGetSurfRef medium group=cuModuleGetSurfRef variants=cuModuleGetSurfRef; before=0x179980,0x1ebb60,0x2baed0,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1ebb60,0x179980; outside_wrappers=0 unverified +0x001ec6e0 impl_cuParamSetSize high group=cuParamSetSize variants=cuParamSetSize; before=0x1ec6e0,0x2bb420,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1ec6e0; outside_wrappers=0 unverified +0x001ec8b0 impl_cuParamSetTexRef high group=cuParamSetTexRef variants=cuParamSetTexRef; before=0x1ec8b0,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50; after=0x1ec8b0; outside_wrappers=0 unverified +0x001eca70 impl_cuOccupancyAvailableDynamicSMemPerBlock high group=cuOccupancyAvailableDynamicSMemPerBlock variants=cuOccupancyAvailableDynamicSMemPerBlock; before=0x1eca70,0x2fe540,0x4a7e90,0xd4dc50; after=0x1eca70; outside_wrappers=0 unverified +0x001ecf80 impl_cuParamSetv high group=cuParamSetv variants=cuParamSetv; before=0x179c00,0x1ecf80,0x2bb420,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1ecf80; outside_wrappers=0 unverified +0x001ed180 impl_cuParamSeti high group=cuParamSeti variants=cuParamSeti; before=0x1ed180,0x2bb420,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1ed180; outside_wrappers=0 unverified +0x001ed340 impl_cuParamSetf high group=cuParamSetf variants=cuParamSetf; before=0x1ed340,0x2bb420,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1ed340; outside_wrappers=0 unverified +0x001ed5c0 impl_cuProfilerStart medium group=cuProfilerStart variants=cuProfilerStart; before=0x1798f0,0x1ed5c0,0x2fe230,0x2fe440,0x2fe540,0x4a7e90,0xd4db70,0xd4dc50; after=0x1ed5c0; outside_wrappers=1 unverified +0x001ed5c0 impl_cuProfilerStop medium group=cuProfilerStop variants=cuProfilerStop; before=0x1798f0,0x1ed5c0,0x2fe230,0x2fe440,0x2fe540,0x4a7e90,0xd4db70,0xd4dc50; after=0x1ed5c0; outside_wrappers=1 unverified +0x001ed720 impl_cuSurfObjectGetResourceDesc high group=cuSurfObjectGetResourceDesc variants=cuSurfObjectGetResourceDesc; before=0x1ed720,0x2d8f40,0x2fe540,0x4a7e90,0xd4db70,0xd4dc50,0xd51980; after=0x1ed720; outside_wrappers=0 unverified +0x001ed850 impl_cuTexObjectGetResourceDesc high group=cuTexObjectGetResourceDesc variants=cuTexObjectGetResourceDesc; before=0x1ed850,0x2d8f40,0x2fe540,0x4a7e90,0xd4db70,0xd4dc50,0xd51980; after=0x1ed850; outside_wrappers=0 unverified +0x001eda80 impl_cuTexObjectGetTextureDesc high group=cuTexObjectGetTextureDesc variants=cuTexObjectGetTextureDesc; before=0x1eda80,0x2d8f40,0x2fe540,0x4a7e90,0xd4db70,0xd4dc50,0xd51980; after=0x1eda80; outside_wrappers=0 unverified +0x001edc80 impl_cuTexObjectGetResourceViewDesc high group=cuTexObjectGetResourceViewDesc variants=cuTexObjectGetResourceViewDesc; before=0x1edc80,0x2d8f40,0x2fe540,0x4a7e90,0xd4db70,0xd4dc50,0xd51980; after=0x1edc80; outside_wrappers=0 unverified +0x001f0fd0 impl_cuSurfObjectDestroy medium group=cuSurfObjectDestroy variants=cuSurfObjectDestroy; before=0x1f0fd0,0x2d8f40,0x2f9890,0x2fe230,0x2fe540,0x4a7e90,0xd4db70,0xd4dc50,0xd51980; after=0x1f0fd0,0x2f9890; outside_wrappers=0 unverified +0x001f11d0 impl_cuSurfRefSetArray high group=cuSurfRefSetArray variants=cuSurfRefSetArray; before=0x1f11d0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1f11d0; outside_wrappers=0 unverified +0x001f1540 impl_cuSurfObjectCreate high group=cuSurfObjectCreate variants=cuSurfObjectCreate; before=0x1f1540,0x2fe540,0x4a7e90,0xd4dc50; after=0x1f1540; outside_wrappers=0 unverified +0x001f2970 impl_cuTexObjectCreate high group=cuTexObjectCreate variants=cuTexObjectCreate; before=0x1f2970,0x2fe540,0x4a7e90,0xd4dc50; after=0x1f2970; outside_wrappers=0 unverified +0x001f35d0 impl_cuTexObjectDestroy medium group=cuTexObjectDestroy variants=cuTexObjectDestroy; before=0x1f35d0,0x2d8f40,0x2fbad0,0x2fe230,0x2fe540,0x4a6e10,0x4a7e90,0xd4db70,0xd4dc50,0xd51980; after=0x1f35d0,0x4a6e10,0x2fbad0; outside_wrappers=0 unverified +0x001f3830 impl_cuTexRefDestroy medium group=cuTexRefDestroy variants=cuTexRefDestroy; before=0x1798f0,0x1f3830,0x2d8f40,0x2fbad0,0x2fe440,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f3830,0x2fbad0; outside_wrappers=0 unverified +0x001f3990 impl_cuTexRefSetArray high group=cuTexRefSetArray variants=cuTexRefSetArray; before=0x1f3990,0x2fe540,0x4a7e90,0xd4dc50; after=0x1f3990; outside_wrappers=0 unverified +0x001f3ca0 impl_cuTexRefSetMipmappedArray high group=cuTexRefSetMipmappedArray variants=cuTexRefSetMipmappedArray; before=0x1f3ca0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1f3ca0; outside_wrappers=0 unverified +0x001f4000 impl_cuTexRefSetFormat high group=cuTexRefSetFormat variants=cuTexRefSetFormat; before=0x1f4000,0x2583d0,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f4000; outside_wrappers=0 unverified +0x001f4200 impl_cuTexRefSetAddressMode high group=cuTexRefSetAddressMode variants=cuTexRefSetAddressMode; before=0x1f4200,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f4200; outside_wrappers=0 unverified +0x001f4370 impl_cuTexRefSetFilterMode high group=cuTexRefSetFilterMode variants=cuTexRefSetFilterMode; before=0x1f4370,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f4370; outside_wrappers=0 unverified +0x001f44b0 impl_cuTexRefSetBorderColor high group=cuTexRefSetBorderColor variants=cuTexRefSetBorderColor; before=0x1f44b0,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f44b0; outside_wrappers=0 unverified +0x001f4620 impl_cuTexRefSetFlags high group=cuTexRefSetFlags variants=cuTexRefSetFlags; before=0x1f4620,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f4620; outside_wrappers=0 unverified +0x001f4770 impl_cuTexRefSetMipmapFilterMode high group=cuTexRefSetMipmapFilterMode variants=cuTexRefSetMipmapFilterMode; before=0x1f4770,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f4770; outside_wrappers=0 unverified +0x001f48b0 impl_cuTexRefSetMipmapLevelBias high group=cuTexRefSetMipmapLevelBias variants=cuTexRefSetMipmapLevelBias; before=0x1f48b0,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f48b0; outside_wrappers=0 unverified +0x001f49f0 impl_cuTexRefSetMipmapLevelClamp high group=cuTexRefSetMipmapLevelClamp variants=cuTexRefSetMipmapLevelClamp; before=0x1f49f0,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f49f0; outside_wrappers=0 unverified +0x001f4b60 impl_cuTexRefSetMaxAnisotropy high group=cuTexRefSetMaxAnisotropy variants=cuTexRefSetMaxAnisotropy; before=0x1f4b60,0x2d8f40,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50,0xd51980; after=0x1f4b60; outside_wrappers=0 unverified +0x001f4f60 impl_cuTexRefGetAddressMode high group=cuTexRefGetAddressMode variants=cuTexRefGetAddressMode; before=0x1798f0,0x1f4f60,0x2fe440,0x2fe540,0x4a7e90,0xd11b90,0xd4dc50; after=0x1f4f60; outside_wrappers=0 unverified +0x001f5c20 impl_cuGraphicsVDPAURegisterVideoSurface high group=cuGraphicsVDPAURegisterVideoSurface variants=cuGraphicsVDPAURegisterVideoSurface; before=0x179200,0x1799e0,0x1f5c20,0x2d8f40,0x2fe230,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1f5c20; outside_wrappers=0 unverified +0x001f5e10 impl_cuGraphicsVDPAURegisterOutputSurface high group=cuGraphicsVDPAURegisterOutputSurface variants=cuGraphicsVDPAURegisterOutputSurface; before=0x179200,0x1799e0,0x1f5e10,0x2d8f40,0x2fe230,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x1f5e10; outside_wrappers=0 unverified +0x001f8c90 impl_cuMemPoolDestroy medium group=cuMemPoolDestroy variants=cuMemPoolDestroy; before=0x1798f0,0x1d5910,0x1f8c90,0x2fe230,0x2fe440,0x2fe540,0x4a7e90,0xd4dc50; after=0x1f8c90,0x1d5910; outside_wrappers=0 unverified +0x001f9780 impl_cuMulticastUnbind high group=cuMulticastUnbind variants=cuMulticastUnbind; before=0x1f94c0,0x1f9780,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1f9780; outside_wrappers=0 unverified +0x001f9950 impl_cuMulticastAddDevice high group=cuMulticastAddDevice variants=cuMulticastAddDevice; before=0x1f94c0,0x1f9950,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1f9950; outside_wrappers=0 unverified +0x001f9a90 impl_cuMulticastGetGranularity high group=cuMulticastGetGranularity variants=cuMulticastGetGranularity; before=0x1f94c0,0x1f9a90,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1f9a90; outside_wrappers=0 unverified +0x001f9f70 impl_cuOccupancyMaxActiveBlocksPerMultiprocessor medium group=cuOccupancyMaxActiveBlocksPerMultiprocessor variants=cuOccupancyMaxActiveBlocksPerMultiprocessor; before=0x1f9e20,0x1f9f70,0x2bb420,0x2dc100,0x2e4340,0x2fe540,0x4a7e90,0xd4bd20,0xd4dc50; after=0x1f9f70,0x2e4340; outside_wrappers=1 unverified +0x001f9f70 impl_cuOccupancyMaxActiveBlocksPerMultiprocessorWithFlags medium group=cuOccupancyMaxActiveBlocksPerMultiprocessorWithFlags variants=cuOccupancyMaxActiveBlocksPerMultiprocessorWithFlags; before=0x1f9e20,0x1f9f70,0x2bb420,0x2dc100,0x2e4340,0x2fe540,0x4a7e90,0xd4bd20,0xd4dc50; after=0x1f9f70,0x2e4340; outside_wrappers=1 unverified +0x001fa330 impl_cuOccupancyMaxPotentialBlockSize medium group=cuOccupancyMaxPotentialBlockSize variants=cuOccupancyMaxPotentialBlockSize; before=0x1f9e20,0x1fa330,0x2baed0,0x2d8f40,0x2e4570,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1fa330,0x2e4570; outside_wrappers=1 unverified +0x001fa330 impl_cuOccupancyMaxPotentialBlockSizeWithFlags medium group=cuOccupancyMaxPotentialBlockSizeWithFlags variants=cuOccupancyMaxPotentialBlockSizeWithFlags; before=0x1f9e20,0x1fa330,0x2baed0,0x2d8f40,0x2e4570,0x2fde50,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x1fa330,0x2e4570; outside_wrappers=1 unverified +0x001fa6d0 impl_cuOccupancyMaxPotentialClusterSize high group=cuOccupancyMaxPotentialClusterSize variants=cuOccupancyMaxPotentialClusterSize; before=0x1fa6d0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1fa6d0; outside_wrappers=0 unverified +0x001faa00 impl_cuOccupancyMaxActiveClusters high group=cuOccupancyMaxActiveClusters variants=cuOccupancyMaxActiveClusters; before=0x1faa00,0x2fe540,0x4a7e90,0xd4dc50; after=0x1faa00; outside_wrappers=0 unverified +0x001facd0 impl_cuStreamCreate medium group=cuStreamCreate variants=cuStreamCreate; before=0x1facd0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1facd0; outside_wrappers=1 unverified +0x001facd0 impl_cuStreamCreateWithPriority medium group=cuStreamCreateWithPriority variants=cuStreamCreateWithPriority; before=0x1facd0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1facd0; outside_wrappers=1 unverified +0x001fb3d0 impl_cuStreamWaitEvent high group=cuStreamWaitEvent variants=cuStreamWaitEvent_ptsz,cuStreamWaitEvent; before=0x1792f0,0x179350,0x1fb3d0,0x2a2b90,0x2a50f0,0x2baed0,0x2ed7d0,0x2fe540,0x3abad0,0x4a7e90,0xd1d500,0xd1e090,0xd4dc50; after=0x1fb3d0; outside_wrappers=0 unverified +0x001fb7f0 impl_cuStreamGetCaptureInfo high group=cuStreamGetCaptureInfo variants=cuStreamGetCaptureInfo_v2_ptsz,cuStreamGetCaptureInfo_v3_ptsz,cuStreamGetCaptureInfo,cuStreamGetCaptureInfo_ptsz,cuStreamGetCaptureInfo_v2,cuStreamGetCaptureInfo_v3; before=0x1fb7f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1fb7f0; outside_wrappers=0 unverified +0x001fbe30 impl_cuStreamBeginCaptureToGraph high group=cuStreamBeginCaptureToGraph variants=cuStreamBeginCaptureToGraph_ptsz,cuStreamBeginCaptureToGraph; before=0x1fba70,0x1fbe30,0x2d8f40,0x2eff60,0x2f16c0,0x2fe540,0x4a7e90,0xd1e090,0xd4dc50,0xd51980; after=0x1fbe30; outside_wrappers=0 unverified +0x001fd190 impl_cuStreamCopyAttributes high group=cuStreamCopyAttributes variants=cuStreamCopyAttributes_ptsz,cuStreamCopyAttributes; before=0x1fd190,0x2fe540,0x4a7e90,0xd1d500,0xd4dc50; after=0x1fd190; outside_wrappers=0 unverified +0x001fd2e0 impl_cuStreamGetAttribute high group=cuStreamGetAttribute variants=cuStreamGetAttribute_ptsz,cuStreamGetAttribute; before=0x1fd2e0,0x2d8f40,0x2fe540,0x4a7e90,0xd1d500,0xd4dc50,0xd51980; after=0x1fd2e0; outside_wrappers=0 unverified +0x001fd460 impl_cuVDPAUCtxCreate high group=cuVDPAUCtxCreate variants=cuVDPAUCtxCreate_v2,cuVDPAUCtxCreate; before=0x1fd460,0x2689e0,0x26d8e0,0x277680,0x2fe020,0x2fe540,0x4a7e90,0xd4dc50; after=0x1fd460; outside_wrappers=0 unverified +0x001ff270 impl_cuMulticastBindMem medium group=cuMulticastBindMem variants=cuMulticastBindMem; before=0x1f94c0,0x1ff270,0x2fde40,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x1ff270,0x2fde40; outside_wrappers=0 unverified +0x001ff3e0 impl_cuMulticastBindAddr high group=cuMulticastBindAddr variants=cuMulticastBindAddr; before=0x1ff3e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x1ff3e0; outside_wrappers=0 unverified +0x0024a6d0 impl_cuCoredumpSetAttributeGlobal high group=cuCoredumpSetAttributeGlobal variants=cuCoredumpSetAttributeGlobal; before=0x24a4d0,0x24a6d0,0x2fe540,0x4a7e90,0xd4dc50; after=0x24a6d0; outside_wrappers=0 unverified +0x00259480 impl_cuTexRefSetAddress2D high group=cuTexRefSetAddress2D variants=cuTexRefSetAddress2D_v3,cuTexRefSetAddress2D_v2,cuTexRefSetAddress2D; before=0x259480,0x2baed0,0x2d8f40,0x2fbe80,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x259480; outside_wrappers=0 unverified +0x00261f70 impl_cuCheckpointProcessGetRestoreThreadId high group=cuCheckpointProcessGetRestoreThreadId variants=cuCheckpointProcessGetRestoreThreadId; before=0x261f70,0x2fe540,0x4a7e90,0xd4dc50; after=0x261f70; outside_wrappers=0 unverified +0x00262690 impl_cuCheckpointProcessGetState medium group=cuCheckpointProcessGetState variants=cuCheckpointProcessGetState; before=0x179200,0x2623e0,0x262690,0x2fe540,0x30a870,0x4a7e90,0xd4dc50; after=0x262690,0x2623e0,0x30a870; outside_wrappers=0 unverified +0x00266fc0 impl_cuCoredumpSetAttribute high group=cuCoredumpSetAttribute variants=cuCoredumpSetAttribute; before=0x266fc0,0x2fe540,0x4a7e90,0xd4dc50; after=0x266fc0; outside_wrappers=0 unverified +0x0026a9b0 impl_cuCtxSynchronize medium group=cuCtxSynchronize variants=cuCtxSynchronize; before=0x1ad290,0x26a9b0,0x2baed0,0x2fde50,0x2fe540,0x4a7e90,0xd1e3f0,0xd4dc50; after=0x26a9b0,0x1ad290; outside_wrappers=0 unverified +0x0026ec30 impl_cuEGLStreamProducerReturnFrame medium group=cuEGLStreamProducerReturnFrame variants=cuEGLStreamProducerReturnFrame; before=0x1b0db0,0x26ec30,0x2fde50,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50,0xd6e3d0; after=0x26ec30,0x1b0db0,0xd6e3d0; outside_wrappers=0 unverified +0x0026fa30 impl_cuEGLStreamConsumerReleaseFrame medium group=cuEGLStreamConsumerReleaseFrame variants=cuEGLStreamConsumerReleaseFrame; before=0x1b07f0,0x26e7c0,0x26fa30,0x2fde50,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x26fa30,0x26e7c0,0x1b07f0; outside_wrappers=0 unverified +0x00287bb0 impl_cuGetErrorName medium group=cuGetErrorName variants=cuGetErrorName; before=0x287bb0,0x2fe540,0x4a7e90,0xd4dc50; after=0x287bb0; outside_wrappers=1 unverified +0x00287bb0 impl_cuGetErrorString medium group=cuGetErrorString variants=cuGetErrorString; before=0x287bb0,0x2fe540,0x4a7e90,0xd4dc50; after=0x287bb0; outside_wrappers=1 unverified +0x00287cb0 impl_cuEventElapsedTime medium group=cuEventElapsedTime variants=cuEventElapsedTime,cuEventElapsedTime_v2; before=0x287cb0,0x288ba0,0x2bb140,0x2fe540,0x487640,0x4a7e90,0xd4dc50; after=0x287cb0,0x487640,0x288ba0; outside_wrappers=0 unverified +0x00287ed0 impl_cuEventCreate medium group=cuEventCreate variants=cuEventCreate; before=0x1b0fb0,0x287ed0,0x2bb420,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x287ed0,0x1b0fb0; outside_wrappers=0 unverified +0x00288cd0 impl_cuEventSynchronize medium group=cuEventSynchronize variants=cuEventSynchronize; before=0x1b2600,0x288cd0,0x2bb140,0x2fe230,0x2fe540,0x3034c0,0x4a7e90,0xd1e090,0xd4dc50; after=0x288cd0,0x1b2600,0x3034c0; outside_wrappers=0 unverified +0x002891d0 impl_cuEventDestroy high group=cuEventDestroy variants=cuEventDestroy_v2,cuEventDestroy; before=0x2891d0,0x2baed0,0x2bb140,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x2891d0; outside_wrappers=0 verified +0x00294db0 impl_cuGLGetDevices medium group=cuGLGetDevices variants=cuGLGetDevices_v2,cuGLGetDevices; before=0x294db0,0x29d320,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x294db0,0x29d320; outside_wrappers=0 unverified +0x00294fd0 impl_cuGLInit medium group=cuGLInit variants=cuGLInit; before=0x294fd0,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50; after=0x294fd0; outside_wrappers=1 unverified +0x0029d480 impl_cuGraphCreate high group=cuGraphCreate variants=cuGraphCreate; before=0x29d480,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x29d480; outside_wrappers=0 unverified +0x002a1b10 impl_cuGraphAddMemAllocNode medium group=cuGraphAddMemAllocNode variants=cuGraphAddMemAllocNode; before=0x1baa70,0x2a1930,0x2a1b10,0x2a5430,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2a1b10,0x2a5430,0x1baa70,0x2a1930; outside_wrappers=0 unverified +0x002a3530 impl_cuGraphDestroy medium group=cuGraphDestroy variants=cuGraphDestroy; before=0x2a3530,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2a3530; outside_wrappers=1 unverified +0x002a5a20 impl_cuGraphAddMemFreeNode medium group=cuGraphAddMemFreeNode variants=cuGraphAddMemFreeNode; before=0x1c1db0,0x2a5a20,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2a5a20,0x1c1db0; outside_wrappers=0 unverified +0x002a66e0 impl_cuGraphAddBatchMemOpNode medium group=cuGraphAddBatchMemOpNode variants=cuGraphAddBatchMemOpNode; before=0x1c0740,0x1c0bc0,0x2a66e0,0x2bb140,0x2fe540,0x4a7e90,0xd4dc50; after=0x2a66e0,0x1c0bc0; outside_wrappers=0 unverified +0x002a6bf0 impl_cuGraphDebugDotPrint high group=cuGraphDebugDotPrint variants=cuGraphDebugDotPrint; before=0x2a6bf0,0x2fe540,0x4a7e90,0xd4dc50; after=0x2a6bf0; outside_wrappers=0 verified +0x002aa0a0 impl_cuGraphAddExternalSemaphoresSignalNode medium group=cuGraphAddExternalSemaphoresSignalNode variants=cuGraphAddExternalSemaphoresSignalNode; before=0x1b9cc0,0x2aa0a0,0x2fe540,0x4a7e90,0xd4dc50; after=0x2aa0a0,0x1b9cc0; outside_wrappers=0 unverified +0x002aa2d0 impl_cuGraphAddExternalSemaphoresWaitNode medium group=cuGraphAddExternalSemaphoresWaitNode variants=cuGraphAddExternalSemaphoresWaitNode; before=0x1ba0d0,0x2aa2d0,0x2fe540,0x4a7e90,0xd4dc50; after=0x2aa2d0,0x1ba0d0; outside_wrappers=0 unverified +0x002aabc0 impl_cuGraphClone medium group=cuGraphClone variants=cuGraphClone; before=0x1ba8d0,0x2aabc0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2aabc0,0x1ba8d0; outside_wrappers=0 unverified +0x002ac120 impl_cuGraphExecBatchMemOpNodeSetParams medium group=cuGraphExecBatchMemOpNodeSetParams variants=cuGraphExecBatchMemOpNodeSetParams; before=0x179200,0x179a80,0x179c00,0x1c0740,0x1c0e20,0x2ac120,0x2baed0,0x2fe540,0x4a7e90,0xd183b0,0xd4dc50; after=0x2ac120,0x1c0e20; outside_wrappers=0 unverified +0x002aca10 impl_cuGraphKernelNodeSetParams high group=cuGraphKernelNodeSetParams variants=cuGraphKernelNodeSetParams_v2,cuGraphKernelNodeSetParams; before=0x1c2b10,0x2aca10,0x2fe540,0x4a7e90,0xd4dc50; after=0x2aca10; outside_wrappers=0 unverified +0x002afcf0 impl_cuGraphNodeSetEnabled high group=cuGraphNodeSetEnabled variants=cuGraphNodeSetEnabled; before=0x2afcf0,0x2fe540,0x4a7e90,0xd183b0,0xd4dc50; after=0x2afcf0; outside_wrappers=0 unverified +0x002b01c0 impl_cuGraphExecKernelNodeSetParams medium group=cuGraphExecKernelNodeSetParams variants=cuGraphExecKernelNodeSetParams_v2,cuGraphExecKernelNodeSetParams; before=0x1c2b10,0x2b01c0,0x2fe540,0x4a7e90,0xd183b0,0xd4dc50,0xd6a9f0; after=0x2b01c0,0xd6a9f0; outside_wrappers=0 unverified +0x002b4f80 impl_cuGraphExecChildGraphNodeSetParams medium group=cuGraphExecChildGraphNodeSetParams variants=cuGraphExecChildGraphNodeSetParams; before=0x2ac1b0,0x2b4f80,0x2fe540,0x4a7e90,0xd183b0,0xd4dc50; after=0x2b4f80,0x2ac1b0; outside_wrappers=1 unverified +0x002b9ff0 impl_cuGreenCtxDestroy medium group=cuGreenCtxDestroy variants=cuGreenCtxDestroy; before=0x1792f0,0x179350,0x1c6be0,0x26d680,0x2b9ff0,0x2bb140,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2b9ff0,0x1c6be0; outside_wrappers=0 unverified +0x002ba3e0 impl_cuGreenCtxCreate medium group=cuGreenCtxCreate variants=cuGreenCtxCreate; before=0x1792f0,0x179350,0x1c6830,0x26d8a0,0x277570,0x2ba3e0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2ba3e0,0x1c6830,0x277570,0x26d8a0; outside_wrappers=0 unverified +0x002bdf80 impl_cuLinkDestroy high group=cuLinkDestroy variants=cuLinkDestroy; before=0x2bdf80,0x2fe540,0x4a7e90,0xd4dc50; after=0x2bdf80; outside_wrappers=0 unverified +0x002bf2f0 impl_cuKernelGetAttribute high group=cuKernelGetAttribute variants=cuKernelGetAttribute; before=0x2bb420,0x2bf2f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x2bf2f0; outside_wrappers=0 unverified +0x002bfd50 impl_cuKernelSetAttribute medium group=cuKernelSetAttribute variants=cuKernelSetAttribute; before=0x179730,0x1ca190,0x2bb420,0x2bfd50,0x2d8f40,0x2fe540,0x4a7e90,0xd4dc50,0xd51f50; after=0x2bfd50,0x1ca190,0xd51f50,0x179730; outside_wrappers=0 unverified +0x002c05e0 impl_cuCtxSetLimit medium group=cuCtxSetLimit variants=cuCtxSetLimit; before=0x1ca690,0x2c05e0,0x2d8f40,0x2fe230,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x2c05e0,0x1ca690; outside_wrappers=0 unverified +0x002c1040 impl_cuDeviceGetP2PAttribute medium group=cuDeviceGetP2PAttribute variants=cuDeviceGetP2PAttribute; before=0x1af7e0,0x2c1040,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2c1040,0x1af7e0; outside_wrappers=0 unverified +0x002c2890 impl_cuLibraryUnload medium group=cuLibraryUnload variants=cuLibraryUnload; before=0x1c9e80,0x2c2890,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x2c2890,0x1c9e80; outside_wrappers=0 unverified +0x002c4050 impl_cuMemPoolCreate medium group=cuMemPoolCreate variants=cuMemPoolCreate; before=0x1d5780,0x2c4050,0x2c6030,0x2cee80,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50; after=0x2c4050,0x1d5780,0x2cee80,0x2c6030; outside_wrappers=0 unverified +0x002c71c0 impl_cuDeviceGraphMemTrim high group=cuDeviceGraphMemTrim variants=cuDeviceGraphMemTrim; before=0x2c71c0,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2c71c0; outside_wrappers=0 unverified +0x002cb940 impl_cuInit medium group=cuInit variants=cuInit; before=0x2ba8f0,0x2cb940,0x2fe540,0x4a7e90,0xd4dc50; after=0x2cb940,0x2ba8f0; outside_wrappers=0 unverified +0x002ce4e0 impl_cuLinkCreate high group=cuLinkCreate variants=cuLinkCreate_v2,cuLinkCreate; before=0x2bb420,0x2bbd30,0x2ce4e0,0x2fe540,0x4a7e90,0xd4dc50; after=0x2ce4e0; outside_wrappers=0 verified +0x002d09f0 impl_cuLinkComplete high group=cuLinkComplete variants=cuLinkComplete; before=0x2d09f0,0x2fe540,0x4a7e90,0xd4dc50; after=0x2d09f0; outside_wrappers=0 unverified +0x002d3f60 impl_cuLibraryLoadData medium group=cuLibraryLoadData variants=cuLibraryLoadData; before=0x2bbd30,0x2d3f60,0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x2d3f60; outside_wrappers=1 unverified +0x002d81b0 impl_cuMemBatchDecompressAsync medium group=cuMemBatchDecompressAsync variants=cuMemBatchDecompressAsync_ptsz,cuMemBatchDecompressAsync; before=0x26bc10,0x2d8060,0x2d81b0,0x2f8e10,0x2f8fd0,0x2fe540,0x4a7e90,0xd1d500,0xd4dc50,0xd73560; after=0x2d81b0,0xd73560,0x2d8060; outside_wrappers=0 unverified +0x002e84d0 impl_cuModuleUnload medium group=cuModuleUnload variants=cuModuleUnload; before=0x1eb2e0,0x2baed0,0x2d8f40,0x2e84d0,0x2fe230,0x2fe540,0x301460,0x4a7e90,0xd4dc50,0xd4f3e0,0xd51980; after=0x2e84d0,0x1eb2e0,0x301460; outside_wrappers=0 unverified +0x002eb720 impl_cuMemMapArrayAsync medium group=cuMemMapArrayAsync variants=cuMemMapArrayAsync_ptsz,cuMemMapArrayAsync; before=0x1dfae0,0x2d8f40,0x2eb720,0x2fe540,0x4a7e90,0xd1d500,0xd4dc50,0xd51980; after=0x2eb720,0x1dfae0; outside_wrappers=0 unverified +0x002ec4a0 impl_cuStreamAttachMemAsync medium group=cuStreamAttachMemAsync variants=cuStreamAttachMemAsync_ptsz,cuStreamAttachMemAsync; before=0x1fce40,0x2d8f40,0x2ec4a0,0x2f16c0,0x2fe230,0x2fe540,0x4a7e90,0xd4dc50,0xd51980; after=0x2ec4a0,0x1fce40; outside_wrappers=0 unverified +0x002f0110 impl_cuStreamEndCapture medium group=cuStreamEndCapture variants=cuStreamEndCapture_ptsz,cuStreamEndCapture; before=0x1fb6b0,0x2f0110,0x2f16c0,0x2fe540,0x4a7e90,0xd1e090,0xd4dc50; after=0x2f0110,0x1fb6b0; outside_wrappers=0 unverified +0x002f0630 impl_cuStreamDestroy medium group=cuStreamDestroy variants=cuStreamDestroy_v2,cuStreamDestroy; before=0x1792f0,0x179350,0x1fb130,0x2d8f40,0x2f0630,0x2fe540,0x4a7e90,0xd1d500,0xd4dc50,0xd51980; after=0x2f0630,0x1fb130; outside_wrappers=0 unverified +0x002f0fe0 impl_cuGLUnmapBufferObject medium group=cuGLUnmapBufferObject variants=cuGLUnmapBufferObject; before=0x1c1a20,0x2baed0,0x2f0fe0,0x2fde50,0x2fe230,0x2fe540,0x3aa550,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50,0xd663e0; after=0x2f0fe0,0x1c1a20; outside_wrappers=1 unverified +0x002f0fe0 impl_cuGLUnmapBufferObjectAsync medium group=cuGLUnmapBufferObjectAsync variants=cuGLUnmapBufferObjectAsync; before=0x1c1a20,0x2baed0,0x2f0fe0,0x2fde50,0x2fe230,0x2fe540,0x3aa550,0x3c52d0,0x3c5370,0x4a7e90,0xd4dc50,0xd663e0; after=0x2f0fe0,0x1c1a20; outside_wrappers=1 unverified +0x002f1880 impl_cuStreamSetAttribute high group=cuStreamSetAttribute variants=cuStreamSetAttribute_ptsz,cuStreamSetAttribute; before=0x2d8f40,0x2f1880,0x2fe540,0x4a7e90,0xd1d500,0xd4dc50,0xd51980; after=0x2f1880; outside_wrappers=0 unverified +0x002f9f00 impl_cuModuleGetGlobal high group=cuModuleGetGlobal variants=cuModuleGetGlobal_v2,cuModuleGetGlobal; before=0x2baed0,0x2d8f40,0x2f9f00,0x2fe540,0x402500,0x4a7e90,0xd4dc50,0xd51980; after=0x2f9f00; outside_wrappers=1 verified +0x002fb8e0 impl_cuTexRefCreate medium group=cuTexRefCreate variants=cuTexRefCreate; before=0x1f2870,0x2d8f40,0x2fb8e0,0x2fe540,0x4a7e90,0xd4da90,0xd4dc50,0xd51980; after=0x2fb8e0,0x1f2870; outside_wrappers=0 unverified +0x002fffa0 impl_cuMemRangeGetAttribute medium group=cuMemRangeGetAttribute variants=cuMemRangeGetAttribute; before=0x1cb030,0x2fe540,0x2fffa0,0x306160,0x3fe700,0x4a7e90,0xd4bd90,0xd4db70,0xd4dc50; after=0x2fffa0,0x1cb030; outside_wrappers=0 unverified +0x00304420 impl_cuStreamQuery medium group=cuStreamQuery variants=cuStreamQuery_ptsz,cuStreamQuery; before=0x1fcfe0,0x2ea920,0x2f16c0,0x2fe230,0x2fe540,0x304420,0x4a7e90,0xd4dc50; after=0x304420,0x1fcfe0,0x2ea920; outside_wrappers=0 unverified +0x003aa910 impl_cuLibraryEnumerateKernels medium group=cuLibraryEnumerateKernels variants=cuLibraryEnumerateKernels; before=0x2fe540,0x3aa910,0x3aad50,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x3aa910,0x3aad50; outside_wrappers=1 unverified +0x003aa910 impl_cuLibraryGetKernelCount medium group=cuLibraryGetKernelCount variants=cuLibraryGetKernelCount; before=0x2fe540,0x3aa910,0x3aad50,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x3aa910,0x3aad50; outside_wrappers=1 unverified +0x003c9250 impl_cuMemFreeAsync high group=cuMemFreeAsync variants=cuMemFreeAsync_ptsz,cuMemFreeAsync; before=0x1798f0,0x2fe440,0x2fe540,0x3c9250,0x4a7e90,0xd4dc50; after=0x3c9250; outside_wrappers=0 verified +0x003fb7b0 impl_cuMemRelease medium group=cuMemRelease variants=cuMemRelease; before=0x1792f0,0x179350,0x1d2dc0,0x2fe540,0x3fb7b0,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x3fb7b0,0x1d2dc0; outside_wrappers=0 unverified +0x003fcfa0 impl_cuMulticastCreate medium group=cuMulticastCreate variants=cuMulticastCreate; before=0x1792f0,0x179350,0x1f94c0,0x1f9580,0x26bc10,0x2fe540,0x3fcfa0,0x4a7e90,0xd4d9a0,0xd4dc50; after=0x3fcfa0,0x1f9580; outside_wrappers=0 unverified +0x00400ec0 impl_cuMemAlloc high group=cuMemAlloc variants=cuMemAlloc_v2,cuMemAlloc; before=0x1d9cb0,0x2bb300,0x2c1d40,0x2d8f40,0x2e7bf0,0x2fe230,0x2fe540,0x400550,0x400ec0,0x4a7e90,0xd4dc50,0xd51980; after=0x400ec0; outside_wrappers=0 verified +0x00402a30 impl_cuMemGetAddressRange high group=cuMemGetAddressRange variants=cuMemGetAddressRange_v2,cuMemGetAddressRange; before=0x1792f0,0x179350,0x2bb300,0x2fe540,0x3ac3f0,0x402a30,0x403470,0x4a7e90,0xd4bd90,0xd4dc50; after=0x402a30; outside_wrappers=0 verified +0x00d11d90 impl_cuStreamBeginCapture high group=cuStreamBeginCapture variants=cuStreamBeginCapture_v2_ptsz,cuStreamBeginCapture,cuStreamBeginCapture_ptsz,cuStreamBeginCapture_v2; before=0x2d8f40,0x2eff60,0x2f16c0,0x2fe540,0x4a7e90,0xd11d90,0xd4dc50,0xd51980; after=0xd11d90; outside_wrappers=0 verified +0x00d11e90 impl_cuStreamSynchronize medium group=cuStreamSynchronize variants=cuStreamSynchronize_ptsz,cuStreamSynchronize; before=0x1fb280,0x2f16c0,0x2fe230,0x2fe540,0x4a7e90,0xd11e90,0xd4dc50; after=0xd11e90,0x1fb280; outside_wrappers=0 unverified +0x00d121d0 impl_cuGetProcAddress high group=cuGetProcAddress variants=cuGetProcAddress_v2,cuGetProcAddress; before=0x2fe540,0x4a7e90,0xd121d0,0xd4dc50; after=0xd121d0; outside_wrappers=0 unverified +0x00d180a0 impl_cuGraphAddChildGraphNode medium group=cuGraphAddChildGraphNode variants=cuGraphAddChildGraphNode; before=0x1ba7c0,0x2fe540,0x4a7e90,0xd180a0,0xd4d9a0,0xd4dc50; after=0xd180a0,0x1ba7c0; outside_wrappers=0 unverified +0x00d20520 impl_cuIpcOpenEventHandle medium group=cuIpcOpenEventHandle variants=cuIpcOpenEventHandle; before=0x179200,0x1799e0,0x1b3530,0x2bb420,0x2d8f40,0x2da1f0,0x2fe540,0x483550,0x4a7e90,0xd20520,0xd4dc50,0xd4e540,0xd51980; after=0xd20520,0x1b3530,0xd4e540,0x483550,0x2da1f0; outside_wrappers=0 unverified +0x00d668f0 impl_cuLibraryGetUnifiedFunction high group=cuLibraryGetUnifiedFunction variants=cuLibraryGetUnifiedFunction; before=0x2fe540,0x4a7e90,0xd4d9a0,0xd4dc50,0xd668f0; after=0xd668f0; outside_wrappers=0 verified diff --git a/userspace/cuda/570.144/annotations/names.tsv b/userspace/cuda/570.144/annotations/names.tsv index 403bef3..e668055 100644 --- a/userspace/cuda/570.144/annotations/names.tsv +++ b/userspace/cuda/570.144/annotations/names.tsv @@ -1302,3 +1302,42 @@ 0x4676f0 cudaEngineContextCreateHook verified 0x481360 rmEngineContextAlloc verified 0x453cc0 cudaEngineContextCleanup verified +# --- engine-record path verified in the 2026-09 type-recovery pass (see docs/05 section 3.7) --- +0x450ff0 cudaDeviceBuildEngineRecords verified +0x462770 cudaRmGetEngines verified +0x44a170 cudaDeviceSupportsClass verified +0x458040 cudaDeviceAbiTableInit verified +0x44ac20 cudaDeviceInitCaps verified +# +# --- wrapper callees verified in the 2026-09 name-candidate pass --- +# Method: tools/name_candidates.py; per-entry evidence and confidence in +# name-candidates.tsv; roles/evidence also recorded in subsystems.tsv. +0x1a36d0 impl_cuCtxGetApiVersion wrapper-callee +0x1a4900 impl_cuArrayGetMemoryRequirements wrapper-callee +0x1a5410 impl_cuArrayGetPlane wrapper-callee +0x1a5820 impl_cuDevicePrimaryCtxGetState wrapper-callee +0x1a6150 impl_cuDeviceRegisterAsyncNotification wrapper-callee +0x1a6430 impl_cuCtxAttach wrapper-callee +0x1a66f0 impl_cuCtxGetCurrent wrapper-callee +0x1a6850 impl_cuCtxGetDevice wrapper-callee +0x1a6960 impl_cuCtxGetFlags wrapper-callee +0x1a6af0 impl_cuFlushGPUDirectRDMAWrites wrapper-callee +0x1a6c50 impl_cuCtxGetExecAffinity wrapper-callee +0x1a7320 impl_cuDeviceGetLuid wrapper-callee +0x1ae3d0 impl_cuDeviceComputeCapability wrapper-callee +0x1ae550 impl_cuDeviceGetProperties wrapper-callee +0x1ae780 impl_cuDeviceGetAttribute wrapper-callee +0x1be900 impl_cuCtxDestroy wrapper-callee +0x1da450 impl_cuMemFree wrapper-callee +0x1da880 impl_cuMemFreeHost wrapper-callee +0x1e9ab0 impl_cuModuleLoad wrapper-callee +0x1e9d80 impl_cuModuleLoadData wrapper-callee +0x2891d0 impl_cuEventDestroy wrapper-callee +0x2a6bf0 impl_cuGraphDebugDotPrint wrapper-callee +0x2ce4e0 impl_cuLinkCreate wrapper-callee +0x2f9f00 impl_cuModuleGetGlobal wrapper-callee +0x3c9250 impl_cuMemFreeAsync wrapper-callee +0x400ec0 impl_cuMemAlloc wrapper-callee +0x402a30 impl_cuMemGetAddressRange wrapper-callee +0xd11d90 impl_cuStreamBeginCapture wrapper-callee +0xd668f0 impl_cuLibraryGetUnifiedFunction wrapper-callee diff --git a/userspace/cuda/570.144/annotations/subsystems.tsv b/userspace/cuda/570.144/annotations/subsystems.tsv index 5a512bd..4d0c423 100644 --- a/userspace/cuda/570.144/annotations/subsystems.tsv +++ b/userspace/cuda/570.144/annotations/subsystems.tsv @@ -25,13 +25,52 @@ # --- engine/context bring-up (the cuCtxCreate path: crash solved, engine topology blocking; see docs/05) --- 0x3a9c20 cudaHandlePoolAlloc engine bitmap handle allocator: lock, scan 64-bit words for first zero bit, return base + index*64 + bit decompiled 0x3a2f50 cudaHandlePoolFree engine clears the handle bit under the pool mutex decompiled -0x453ef0 cudaEngineContextRegisterHelper engine-engines per-engine registration: alloc handle, call create hook, call bind hook, publish engine id + descriptor into context arrays; called twice for engine index 0 (GR + secondary async CE) decompiled+trace -0x45a670 cudaDeviceInitEngineContexts engine-engines walks 0x14-byte engine records (base+0x2fec for GR; +0x300c+(idx-1)*0x14; +0x350c; +0x351c+(idx-0x16)*0x14), checks engine existence via [table+0x4e00], calls the helper; sees only GR + CE0..4 because the GET_ENGINES_V2/classDB filter drops the async CE (see docs/05 section 3.6) decompiled+trace -0x4676f0 cudaEngineContextCreateHook engine-engines hook at [table+0x40]+0x530; selects instance/class from the engine descriptor and calls rmEngineContextAlloc; its saved rbp was zeroed by the alloc reply's 16-byte class-params copy-back (crash solved, see docs/05 section 3.5) decompiled+trace -0x481360 rmEngineContextAlloc rm builds NV_ESC_RM_ALLOC requests per class (0x80 device, 0x5/0x79 per-class table, 0x2080 subdevice array of 32, default generic) and returns small status codes; with a zeroed secondary engine record the class is rewritten to NV01_ROOT_CLIENT (0x41), status 0x19, and the context rolls back (rc=999) decompiled+trace +0x453ef0 cudaEngineContextRegisterHelper engine-engines per-engine registration: idx = record+0 kind (1 normalizes to 0), alloc handle, call create hook with the record, call bind hook, publish engine id at ctx+0x50+idx*4 and the record pointer at ctx[0x9488]+0x88+idx*8; called twice for engine index 0 (GR + secondary async CE) decompiled+trace +0x45a670 cudaDeviceInitEngineContexts engine-engines per-context engine-init hook installed at ctx+0x94b0; selects the 0x14-byte engine record for ctx+0x38 (GR at +0x2fec; copy at +0x300c+(idx-1)*0x14, +0x300c+(idx-2)*0x14 for idx 0xc..0x15; SEC2 at +0x350c; decomp copy at +0x351c+(idx-0x16)*0x14), checks existence via [table+0x4e00](table, record+0xc), calls the helper; for index 0 also registers the secondary record at *(table+0x4af0)+0x75c decompiled+trace +0x4676f0 cudaEngineContextCreateHook engine-engines create hook at [table+0x40]+0x530; reads record kind +0, instance +8, class +0xc; copies instance into RM alloc-param dword 1 only for kinds 2/3/5; passes record+0xc as the alloc class to rmEngineContextAlloc; its saved rbp was zeroed by the alloc reply's 16-byte class-params copy-back (crash solved, see docs/05 section 3.5); the typed pass recovered its 4 parameters and the field logic (run/analysis/typed/fn_0x4676f0.c) decompiled+trace +0x481360 rmEngineContextAlloc rm builds NV_ESC_RM_ALLOC requests per class (0x80 device, 0x5/0x79 per-class table, 0x2080 subdevice array of 32, default generic with the record as params) and returns small status codes; class comes from the engine record, so a zeroed secondary record means class 0 is rewritten to NV01_ROOT_CLIENT (0x41), status 0x19, and the context rolls back (rc=999) decompiled+trace 0x453cc0 cudaEngineContextCleanup engine-engines failure-path cleanup of an engine context; runs on the rolled-back context (rc=999) decompiled +# --- engine record construction (the record identity and the COPY14 blocker; see docs/05 section 3.7) --- +0x450ff0 cudaDeviceBuildEngineRecords engine-engines builds the 0x14-byte per-engine records from the GET_ENGINES_V2 list at device+0x2aa8 (count +0x2fe8): kind 1 GR -> +0x2fec (class device+0xd38), kind 3 copy -> +0x300c+n*0x14 (class +0xd44, caps bit0 = CE_GRCE also seeds the secondary record at *(device+0x4af0)+0x75c), kind 4 SEC2 -> +0x350c (class +0xd4c), kind 5 decomp copy -> +0x351c+n*0x14 (class +0xd50); fields +0 kind, +4 NV2080 engine type, +8 instance, +0xc class, +0x10 caps decompiled+trace +0x462770 cudaRmGetEngines engine-engines RM control AL wrapper for NV2080_CTRL_CMD_GPU_GET_ENGINES_V2: reads the engine list and classifies each raw NV2080 engine type into {kind, type, instance} records at 0x10-byte stride (GR 1..8 -> kind 1; copy 9..0x12 and 0x34..0x3d -> kind 3; decomp copy 0x40..0x53 -> kind 5; SEC2 0x26 -> kind 4; others skipped); called from 0x44ac20 via AL slot 0x1d8 decompiled +0x44a170 cudaDeviceSupportsClass engine-engines device class-list membership callback installed at device+0x4e00 by the ABI table setup (0x458040 and 0x4494d0): returns true when class is in the class list at device+0xd90 (count +0xd98); the walker calls it with (device, record+0xc) before registering an engine decompiled+trace +0x458040 cudaDeviceAbiTableInit engine-engines builds the RM device state and installs the device ABI function table at device+0x4d18.. (engine builder +0x4d88, class callback +0x4e00 = 0x44a170, ctx table setup +0x4ef0, init caps +0x4d38); class list copied from the config block into device+0xd90/+0xd98 decompiled +0x44ac20 cudaDeviceInitCaps engine-engines device init-caps path: calls the AL getEngines slot 0x1d8 (0x462770) to fill device+0x2aa8/+0x2fe8, then the ABI static-info hooks; the engine list it produces is what the record builder consumes decompiled # --- CURRENT BLOCKER (2026-09-23): engine list / classDB consistency (async CE COPY14) --- -0x45a670 cudaDeviceInitEngineContexts engine-engines CURRENT BLOCKER: classDB-filtered engine list has 6 engines in the guest (GR + CE0..4) vs 7 on the host oracle (adds async CE COPY14: engDesc 0x793ceb0e, instance 14, client type 0x38, class 0xc9b5); without a classDB entry the walker reads a zeroed secondary record and the RM alloc is rejected (class 0x41, st 0x19). Record layout above; see docs/05 section 3.6 decompiled+trace +0x45a670 cudaDeviceInitEngineContexts engine-engines CURRENT BLOCKER: the GET_ENGINES_V2 list has 6 engines in the guest (GR + CE0..4) vs 7 on the host oracle (adds async CE COPY14: engDesc 0x793ceb0e, NV2080 type 0x38, instance 14, paired class 0xc9b5); the classDB filter drops COPY14, so no copy record carries CE_GRCE (caps bit0, live guest 0xbc/0xfc) and the secondary record at *(table+0x4af0)+0x75c stays zero (host t=2 cls=0xc5b5); the zero class is rewritten to 0x41 (st 0x19) and the context rolls back. Record field map and builder chain in docs/05 section 3.7 decompiled+trace # --- open questions --- 0x4a7e90 FUN_004a7e90 unknown called by every generated wrapper; likely tool/trace dispatch or thread-data helper todo 0x2cb940 FUN_002cb940 unknown large function (10 KB) reachable from the cuInit wrapper; candidate init implementation todo +# --- wrapper callees verified in the 2026-09 name-candidate pass --- +# Names proposed by tools/name_candidates.py (see name-candidates.tsv), each +# verified by reading its body in the 570.144 corpus; source `wrapper-callee` +# in names.tsv. Evidence: decompiled. +0x1a36d0 impl_cuCtxGetApiVersion api current-context API version getter: resolves the thread context and returns its API-version field decompiled (wrapper-callee) +0x1a4900 impl_cuArrayGetMemoryRequirements mem array memory-requirements getter: refreshes the backing-store state under the context mutex and returns the allocation size decompiled (wrapper-callee) +0x1a5410 impl_cuArrayGetPlane mem array plane getter: validates the plane index and returns the plane handle decompiled (wrapper-callee) +0x1a5820 impl_cuDevicePrimaryCtxGetState api primary-context state getter: reads flags/active state under the device mutex decompiled (wrapper-callee) +0x1a6150 impl_cuDeviceRegisterAsyncNotification api device async-notification registration: allocates and links a notification record under the device mutex decompiled (wrapper-callee) +0x1a6430 impl_cuCtxAttach api context attach: increments the context attach refcount under the context lock decompiled (wrapper-callee) +0x1a66f0 impl_cuCtxGetCurrent api current-context getter: returns the calling thread's context from TLS decompiled (wrapper-callee) +0x1a6850 impl_cuCtxGetDevice api context device getter: resolves the current context and returns its device ordinal decompiled (wrapper-callee) +0x1a6960 impl_cuCtxGetFlags api context flags getter: composes flags from context fields under the context lock decompiled (wrapper-callee) +0x1a6af0 impl_cuFlushGPUDirectRDMAWrites api GPUDirect RDMA write flush: validates args and dispatches through the device op table decompiled (wrapper-callee) +0x1a6c50 impl_cuCtxGetExecAffinity api context exec-affinity getter: returns the affinity value under the context lock decompiled (wrapper-callee) +0x1a7320 impl_cuDeviceGetLuid api device LUID getter: bounds-checks the ordinal and calls the device op at table+0x5180 decompiled (wrapper-callee) +0x1ae3d0 impl_cuDeviceComputeCapability api compute-capability getter: reads device attributes 0x4b/0x4c decompiled (wrapper-callee) +0x1ae550 impl_cuDeviceGetProperties api device properties getter: fills CUDA device properties from the device record decompiled (wrapper-callee) +0x1ae780 impl_cuDeviceGetAttribute api device attribute getter: validates the ordinal and reads one attribute (fast path for attribute 0x20080403) decompiled (wrapper-callee) +0x1be900 impl_cuCtxDestroy api context destroy: validates the handle, takes the context lock and tears the context down decompiled (wrapper-callee) +0x1da450 impl_cuMemFree mem device memory free: resolves the allocation under the context lock and frees it decompiled (wrapper-callee) +0x1da880 impl_cuMemFreeHost mem host memory free: frees pinned host memory (or routes to the device free path) decompiled (wrapper-callee) +0x1e9ab0 impl_cuModuleLoad module module load from file: reads the file and parses it into the context decompiled (wrapper-callee) +0x1e9d80 impl_cuModuleLoadData module module load from memory image: same path as file load with an in-memory image decompiled (wrapper-callee) +0x2891d0 impl_cuEventDestroy sync event destroy: unlinks and frees an event under the context mutex decompiled (wrapper-callee) +0x2a6bf0 impl_cuGraphDebugDotPrint launch graph DOT dump: writes the graph as a DOT debug file decompiled (wrapper-callee) +0x2ce4e0 impl_cuLinkCreate jit linker create: allocates and initializes the linker state decompiled (wrapper-callee) +0x2f9f00 impl_cuModuleGetGlobal module module global getter: resolves a named global through the module symbol table decompiled (wrapper-callee) +0x3c9250 impl_cuMemFreeAsync mem async device memory free: resolves the allocation and frees it on a stream decompiled (wrapper-callee) +0x400ec0 impl_cuMemAlloc mem device memory alloc: validates/derives the allocation descriptor and allocates decompiled (wrapper-callee) +0x402a30 impl_cuMemGetAddressRange mem allocation lookup: resolves the allocation containing an address under the context mutex decompiled (wrapper-callee) +0xd11d90 impl_cuStreamBeginCapture stream stream begin capture: validates the mode and starts capture under the stream lock decompiled (wrapper-callee) +0xd668f0 impl_cuLibraryGetUnifiedFunction module library function getter: resolves a named function through the library symbol table decompiled (wrapper-callee) diff --git a/userspace/cuda/570.144/ghidra/DecompileAt.java b/userspace/cuda/570.144/ghidra/DecompileAt.java index 5092ee9..f1572c7 100644 --- a/userspace/cuda/570.144/ghidra/DecompileAt.java +++ b/userspace/cuda/570.144/ghidra/DecompileAt.java @@ -1,5 +1,18 @@ // SPDX-License-Identifier: BSD-3-Clause // +// ########################################################################### +// # WARNING: DEPRECATED - DO NOT USE FOR NEW ANALYSIS (kept for history). # +// # # +// # It treats its argument as a Ghidra program address, but the analysis # +// # convention here is FILE virtual addresses (libcuda+0x...) and Ghidra # +// # loads these binaries at image base 0x100000. Called with a file vaddr # +// # it decompiles the function 0x100000 bytes too low (the bug behind the # +// # old `helper_453ef0.c` / `fn_0x481360.c` artifacts). # +// # # +// # Replacement: tools/ghidra/DecompileAt.java (adds the image base) used # +// # through tools/ghidra_decompile_at.sh. # +// ########################################################################### +// // DecompileAt.java - Ghidra post-script: decompile the function containing a // given address (no auto-analysis required). // diff --git a/userspace/cuda/570.144/scripts/ghidra_analyze.sh b/userspace/cuda/570.144/scripts/ghidra_analyze.sh index b844300..eb8422b 100755 --- a/userspace/cuda/570.144/scripts/ghidra_analyze.sh +++ b/userspace/cuda/570.144/scripts/ghidra_analyze.sh @@ -1,6 +1,21 @@ #!/bin/bash # SPDX-License-Identifier: BSD-3-Clause # +# ########################################################################### +# # WARNING: DEPRECATED - DO NOT USE FOR NEW ANALYSIS (kept for history). # +# # # +# # Two bugs make its output unusable: # +# # 1. It passes FILE virtual addresses to Ghidra, which loads these # +# # binaries at image base 0x100000. Every address is 0x100000 too low, # +# # so the decompiled "functions" are the wrong ones. # +# # 2. `-analysisTimeoutPerFile 0` does not mean "no timeout" in this # +# # Ghidra build: it aborts auto-analysis immediately, producing an # +# # analysed-looking project with no analysis at all. # +# # # +# # Replacement: tools/ghidra_full_analysis.sh (plus # +# # tools/ghidra/ConfigureAnalysis.java). # +# ########################################################################### +# # ghidra_analyze.sh - import a library into a Ghidra project and run the full # auto-analysis once. The project is kept, so later decompilations are cheap: # diff --git a/userspace/cuda/570.144/scripts/ghidra_decompile_project.sh b/userspace/cuda/570.144/scripts/ghidra_decompile_project.sh index 66d7c90..b44b2b2 100755 --- a/userspace/cuda/570.144/scripts/ghidra_decompile_project.sh +++ b/userspace/cuda/570.144/scripts/ghidra_decompile_project.sh @@ -1,6 +1,19 @@ #!/bin/bash # SPDX-License-Identifier: BSD-3-Clause # +# ########################################################################### +# # WARNING: DEPRECATED - DO NOT USE FOR NEW ANALYSIS (kept for history). # +# # # +# # It passes a FILE virtual address straight to 570.144/ghidra/ # +# # DecompileAt.java, but Ghidra loads these binaries at image base # +# # 0x100000: the script resolves (and decompiles) the function 0x100000 # +# # bytes too low. It also depends on a project produced by the broken # +# # ghidra_analyze.sh above. # +# # # +# # Replacement: tools/ghidra_decompile_at.sh, which translates the file # +# # vaddr with tools/ghidra/DecompileAt.java (file-vaddr aware). # +# ########################################################################### +# # ghidra_decompile_project.sh - decompile a function from an already-analyzed # Ghidra project (run scripts/ghidra_analyze.sh once first). No re-import and # no analysis, so this takes seconds. |