Skip to content

Make tutorials build on a DGX Spark w/ CUDA 13.0 #3

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Open
gnurizen wants to merge 1 commit into
base: master
Choose a base branch
from
Open

Make tutorials build on a DGX Spark w/ CUDA 13.0 #3

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 1 addition & 1 deletion activity_trace/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -2,7 +2,7 @@
# Copyright 2011-2013 NVIDIA Corporation. All rights reserved
#
INCLUDES=-I../common

Copilot AI Feb 11, 2026

Copy link

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

NVCC is now derived from $(CUDA_INSTALL_PATH), but this Makefile never defines CUDA_INSTALL_PATH (unlike most other samples). On a clean environment this will resolve to /bin/nvcc and break the build. Add a CUDA_INSTALL_PATH ?= ... default (or compute NVCC ?= nvcc from PATH) to keep the sample self-contained.

Suggested change
INCLUDES=-I../common
INCLUDES=-I../common
CUDA_INSTALL_PATH ?= /usr/local/cuda

Copilot uses AI. Check for mistakes.
NVCC=/usr/local/cuda-12.9/bin/nvcc
NVCC=$(CUDA_INSTALL_PATH)/bin/nvcc

ifndef OS
OS := $(shell uname)
Expand Down
6 changes: 2 additions & 4 deletions activity_trace_async/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -75,10 +75,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

activity_trace_async: activity_trace_async.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
6 changes: 2 additions & 4 deletions autorange_profiling/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -82,10 +82,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

.DEFAULT: all
.PHONY: all
Expand Down
2 changes: 1 addition & 1 deletion autorange_profiling/auto_range_profiling.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -426,7 +426,7 @@ main(
}

CUcontext cuContext;
DRIVER_API_CALL(cuCtxCreate(&cuContext, 0, cuDevice));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&cuContext, 0, cuDevice));

// Get chip name for the CUDA device.
CUpti_Device_GetChipName_Params getChipNameParams = { CUpti_Device_GetChipName_Params_STRUCT_SIZE };
Expand Down
2 changes: 1 addition & 1 deletion autorange_profiling/simplecuda.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -217,7 +217,7 @@ bool runTest(int deviceNum,
DRIVER_API_CALL(cuDeviceGet(&cuDevice, deviceNum));

CUcontext cuContext;
DRIVER_API_CALL(cuCtxCreate(&cuContext, 0, cuDevice));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&cuContext, 0, cuDevice));

CUpti_Profiler_BeginSession_Params beginSessionParams = {CUpti_Profiler_BeginSession_Params_STRUCT_SIZE};
CUpti_Profiler_SetConfig_Params setConfigParams = {CUpti_Profiler_SetConfig_Params_STRUCT_SIZE};
Expand Down
6 changes: 2 additions & 4 deletions callback_event/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -76,10 +76,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

callback_event: callback_event.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ callback_event.$(OBJ) $(LIBS) $(INCLUDES)
Expand Down
2 changes: 1 addition & 1 deletion callback_event/callback_event.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -225,7 +225,7 @@ main(
exit(EXIT_WAIVED);
}

DRIVER_API_CALL(cuCtxCreate(&context, 0, dev));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&context, 0, dev));

// Creating event group for profiling
CUPTI_API_CALL(cuptiEventGroupCreate(context, &cuptiEvent.eventGroup, 0));
Expand Down
6 changes: 2 additions & 4 deletions callback_metric/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -76,10 +76,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

callback_metric: callback_metric.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ callback_metric.$(OBJ) $(LIBS) $(INCLUDES)
Expand Down
2 changes: 1 addition & 1 deletion callback_metric/callback_metric.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -372,7 +372,7 @@ main(
exit(EXIT_WAIVED);
}

DRIVER_API_CALL(cuCtxCreate(&context, 0, device));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&context, 0, device));

// Get the name of the metric to collect
if (argc > 2)
Expand Down
6 changes: 2 additions & 4 deletions callback_profiling/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -82,10 +82,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

.DEFAULT: all
.PHONY: all
Expand Down
6 changes: 2 additions & 4 deletions callback_timestamp/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -76,10 +76,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk


callback_timestamp: callback_timestamp.$(OBJ)
Expand Down
2 changes: 1 addition & 1 deletion callback_timestamp/callback_timestamp.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -282,7 +282,7 @@ main(

DRIVER_API_CALL(cuInit(0));

DRIVER_API_CALL(cuCtxCreate(&context, 0, device));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&context, 0, device));

// Allocate input vectors pHostA and pHostB in host memory.
pHostA = (int *)malloc(size);
Expand Down
6 changes: 2 additions & 4 deletions checkpoint_kernels/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -81,10 +81,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

checkpoint_kernels: checkpoint_kernels.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
2 changes: 1 addition & 1 deletion checkpoint_kernels/checkpoint_kernels.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -69,7 +69,7 @@ main()

// Set up a context for device 0.
RUNTIME_API_CALL(cudaSetDevice(0));
DRIVER_API_CALL(cuCtxCreate(&ctx, 0, 0));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&ctx, 0, 0));

// Allocate host and device arrays and initialize to known values.
float *pDeviceA;
Expand Down
23 changes: 23 additions & 0 deletions common/common.mk
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,23 @@
# Common Makefile configuration for CUPTI samples
# Copyright NVIDIA Corporation

# Detect CUDA version to determine supported SM architectures
# CUDA 13+ removed support for compute capabilities below sm_75
CUDA_VERSION := $(shell $(CUDA_INSTALL_PATH)/bin/nvcc --version 2>/dev/null | grep release | sed 's/.*release //' | sed 's/,.*//' | cut -d. -f1)

Copilot AI Feb 11, 2026

Copy link

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

CUDA_VERSION is derived from $(CUDA_INSTALL_PATH)/bin/nvcc directly. Several sample Makefiles allow overriding NVCC (or quote it), and CUDA_INSTALL_PATH might not match the actual nvcc used. Consider basing detection on $(NVCC) instead (or falling back to nvcc on PATH) so the architecture selection matches the compiler actually invoked.

Suggested change
CUDA_VERSION := $(shell $(CUDA_INSTALL_PATH)/bin/nvcc --version 2>/dev/null | grep release | sed 's/.*release //' | sed 's/,.*//' | cut -d. -f1)
NVCC ?= nvcc
CUDA_VERSION := $(shell $(NVCC) --version 2>/dev/null | grep release | sed 's/.*release //' | sed 's/,.*//' | cut -d. -f1)

Copilot uses AI. Check for mistakes.

# Set default SMS based on CUDA version if not already set
ifndef SMS
ifeq ($(shell test $(CUDA_VERSION) -ge 13 2>/dev/null; echo $$?),0)
# CUDA 13+ only supports sm_75 and above
Comment on lines +4 to +11

Copilot AI Feb 11, 2026

Copy link

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

CUDA version detection relies on Unix utilities (grep, sed, cut, test) in a pipeline. These are unlikely to work in the Windows_NT codepaths where GNU make may be using cmd.exe, which would break any sample Makefile that includes this file. Consider guarding auto-detection behind a non-Windows check and/or providing a simple fallback (require SMS to be set on Windows, or use a more portable parsing approach).

Copilot uses AI. Check for mistakes.
# Note: CUDA 13 uses sm_100, sm_103 instead of sm_101, sm_102
SMS := 75 80 86 87 88 89 90 100 103 110 120
DEFAULT_SM_ARCH := sm_75
else
# CUDA 12 and below support older architectures
SMS := 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
DEFAULT_SM_ARCH := sm_52
endif
endif

Copilot AI Feb 11, 2026

Copy link

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

DEFAULT_SM_ARCH is only set inside the ifndef SMS block. If a user overrides SMS (common when targeting a single GPU), DEFAULT_SM_ARCH remains unset and samples like pc_sampling that use -arch=$(DEFAULT_SM_ARCH) will fail. Define DEFAULT_SM_ARCH independently when it is not provided (e.g., derive it from $(firstword $(SMS))).

Suggested change
# If SMS was provided by the user, ensure DEFAULT_SM_ARCH is also set.
# Derive it from the first entry in SMS when not explicitly specified.
ifndef DEFAULT_SM_ARCH
ifdef SMS
DEFAULT_SM_ARCH := sm_$(firstword $(SMS))
endif
endif

Copilot uses AI. Check for mistakes.
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
10 changes: 10 additions & 0 deletions common/helper_cupti.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -49,6 +49,16 @@ do
} while (0)
#endif

// Compatibility wrapper for cuCtxCreate API change in CUDA 13
// CUDA 13+ requires an additional CUctxCreateParams parameter
#if CUDA_VERSION >= 13000
#define COMPAT_cuCtxCreate(pctx, flags, dev) \
cuCtxCreate(pctx, NULL, flags, dev)
#else
#define COMPAT_cuCtxCreate(pctx, flags, dev) \
cuCtxCreate(pctx, flags, dev)
#endif

#ifndef RUNTIME_API_CALL
#define RUNTIME_API_CALL(apiFunctionCall) \
do \
Expand Down
6 changes: 2 additions & 4 deletions concurrent_profiling/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -82,10 +82,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

.DEFAULT: all
.PHONY: all
Expand Down
2 changes: 1 addition & 1 deletion concurrent_profiling/concurrent_profiling.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -614,7 +614,7 @@ main(
// CUPTI_KernelReplay, CUPTI_UserReplay, or CUPTI_ApplicationReplay.
config.replayMode = CUPTI_UserReplay;
// Either set to a context, or may be NULL if a default context has been created.
DRIVER_API_CALL(cuCtxCreate(&(config.context), 0, device));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&(config.context), 0, device));
// Save this device config.
deviceData[device].config = config;

Expand Down
6 changes: 2 additions & 4 deletions cuda_graphs_trace/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -76,10 +76,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

cuda_graphs_trace: cuda_graphs_trace.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
6 changes: 2 additions & 4 deletions cuda_memory_trace/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -75,10 +75,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

cuda_memory_trace: memory_trace.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
6 changes: 2 additions & 4 deletions cupti_correlation/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -75,10 +75,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

cupti_correlation: cupti_correlation.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
6 changes: 2 additions & 4 deletions cupti_external_correlation/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -75,10 +75,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

cupti_external_correlation: cupti_external_correlation.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
2 changes: 1 addition & 1 deletion cupti_external_correlation/cupti_external_correlation.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -97,7 +97,7 @@ DoVectorAddition() {
// Push external id for the initialization: memory allocation and memcpy operations from host to device.
CUPTI_API_CALL_VERBOSE(cuptiActivityPushExternalCorrelationId(CUPTI_EXTERNAL_CORRELATION_KIND_UNKNOWN, static_cast<uint64_t>(INITIALIZATION_EXTERNAL_ID)));

DRIVER_API_CALL(cuCtxCreate(&context, 0, device));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&context, 0, device));

// Allocate vectors in device memory.
RUNTIME_API_CALL(cudaMalloc((void **)&pDeviceA, size));
Expand Down
6 changes: 2 additions & 4 deletions cupti_metric_properties/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -82,10 +82,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 70 72 75 80 86 87 89 90 100
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

.DEFAULT: all
.PHONY: all
Expand Down
2 changes: 1 addition & 1 deletion cupti_metric_properties/cupti_metric_properties.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -454,7 +454,7 @@ CUptiResult GetCounterAvailabilityImage(std::vector<uint8_t>& counterAvailabilit
{
DRIVER_API_CALL(cuInit(0));
CUcontext ctx;
DRIVER_API_CALL(cuCtxCreate(&ctx, 0, 0));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&ctx, 0, 0));

CUpti_Profiler_GetCounterAvailability_Params getCounterAvailabilityParams = { CUpti_Profiler_GetCounterAvailability_Params_STRUCT_SIZE };
getCounterAvailabilityParams.ctx = ctx;
Expand Down
6 changes: 2 additions & 4 deletions cupti_nvtx/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -77,10 +77,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

cupti_nvtx: cupti_nvtx.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
2 changes: 1 addition & 1 deletion cupti_nvtx/cupti_nvtx.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -114,7 +114,7 @@ DoVectorAddition()
InitializeVector(pHostB, N);
memset(pHostC, 0, size);

DRIVER_API_CALL(cuCtxCreate(&context, 0, device));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&context, 0, device));
nvtxNameCuContextA(context, "CUDA Context");

// Push range "Allocate device memory" on domain "Vector Addition".
Expand Down
6 changes: 2 additions & 4 deletions cupti_nvtx_ext_payload/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -77,10 +77,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 75 80 86 87 89 90 100 103 110 120 121
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

cupti_nvtx_ext_payload: cupti_nvtx_ext_payload.$(OBJ) nvtx_payload_attributes.$(OBJ) nvtx_payload_parser.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ $^ $(LIBS) $(INCLUDES)
Expand Down
2 changes: 1 addition & 1 deletion event_multi_gpu/event_multi_gpu.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -103,7 +103,7 @@ main(
{
RUNTIME_API_CALL(cudaSetDevice(i));

DRIVER_API_CALL(cuCtxCreate(&(context[i]), 0, device[i]));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&(context[i]), 0, device[i]));

DRIVER_API_CALL(cuCtxPopCurrent(&(context[i])));
}
Expand Down
6 changes: 2 additions & 4 deletions event_sampling/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -76,10 +76,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 50 52 53 60 61 62 70 72
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

event_sampling: event_sampling.$(OBJ)
$(NVCC) $(NVCC_COMPILER) $(NVCCFLAGS) -o $@ event_sampling.$(OBJ) $(LIBS) $(INCLUDES)
Expand Down
2 changes: 1 addition & 1 deletion event_sampling/event_sampling.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -311,7 +311,7 @@ main(int argc, char *argv[])
s_pEventName = EVENT_NAME;
}

DRIVER_API_CALL(cuCtxCreate(&context, 0, device));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&context, 0, device));

// Create semaphore.
#ifdef _WIN32
Expand Down
6 changes: 2 additions & 4 deletions nested_range_profiling/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -82,10 +82,8 @@ ifneq ($(TARGET_ARCH), $(HOST_ARCH))
endif
endif

# Gencode arguments
SMS ?= 70 72 75 80 86 87 89 90 100 101 120
# Generate SASS code for each SM architecture listed in $(SMS)
$(foreach sm,$(SMS),$(eval GENCODE_FLAGS += -gencode arch=compute_$(sm),code=sm_$(sm)))
# Include common configuration for SMS architecture detection
include ../common/common.mk

.DEFAULT: all
.PHONY: all
Expand Down
2 changes: 1 addition & 1 deletion nested_range_profiling/nested_range_profiling.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -448,7 +448,7 @@ int main(
}

CUcontext cuContext;
DRIVER_API_CALL(cuCtxCreate(&cuContext, 0, cuDevice));
DRIVER_API_CALL(COMPAT_cuCtxCreate(&cuContext, 0, cuDevice));

// Get chip name for the cuda device.
CUpti_Device_GetChipName_Params getChipNameParams = { CUpti_Device_GetChipName_Params_STRUCT_SIZE };
Expand Down
Loading