From fc8282c5c2ecc3536a189fcf69dc70069c789cfd Mon Sep 17 00:00:00 2001 From: Will MacCormack Date: Thu, 19 Oct 2023 21:05:47 -0400 Subject: [PATCH 1/3] Swap out Homebrew paths and recommend a metal ext --- .vscode/c_cpp_properties.json | 6 +++--- .vscode/extensions.json | 5 +++++ .vscode/tasks.json | 38 +++++++++++++++++------------------ 3 files changed, 27 insertions(+), 22 deletions(-) create mode 100644 .vscode/extensions.json diff --git a/.vscode/c_cpp_properties.json b/.vscode/c_cpp_properties.json index cb173bd..ebf8cc9 100644 --- a/.vscode/c_cpp_properties.json +++ b/.vscode/c_cpp_properties.json @@ -3,15 +3,15 @@ "name": "Mac M1 LLVM OpenMP", "includePath": [ "${workspaceFolder}/**", - "/opt/homebrew/opt/libomp/lib" + "/usr/local/opt/libomp/lib" ], "defines": [], "macFrameworkPath": [], - "compilerPath": "/opt/homebrew/opt/llvm/bin/clang++", + "compilerPath": "/usr/local/opt/llvm/bin/clang++", "cStandard": "c17", "cppStandard": "c++20", "intelliSenseMode": "clang-x64", "compilerArgs": [] }], "version": 4 -} \ No newline at end of file +} diff --git a/.vscode/extensions.json b/.vscode/extensions.json new file mode 100644 index 0000000..132d299 --- /dev/null +++ b/.vscode/extensions.json @@ -0,0 +1,5 @@ +{ + "recommendations": [ + "doublebuffer.metal-shader" + ] +} diff --git a/.vscode/tasks.json b/.vscode/tasks.json index 9adb1f7..a10de11 100644 --- a/.vscode/tasks.json +++ b/.vscode/tasks.json @@ -3,13 +3,13 @@ "tasks": [{ "type": "cppbuild", "label": "Build 01", - "command": "/opt/homebrew/opt/llvm/bin/clang++", + "command": "/usr/local/opt/llvm/bin/clang++", "args": [ "-std=c++17", "-stdlib=libc++", "-O2", // OpenMP includes & configuration - "-L/opt/homebrew/opt/libomp/lib", "-fopenmp", + "-L/usr/local/opt/libomp/lib", "-fopenmp", // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", @@ -28,7 +28,7 @@ "kind": "build", "isDefault": true }, - "detail": "compiler: /opt/homebrew/opt/llvm/bin/clang++", + "detail": "compiler: /usr/local/opt/llvm/bin/clang++", "dependsOn": ["Build .metallib 01"], }, // See: https://developer.apple.com/documentation/metal/shader_libraries/building_a_library_with_metal_s_command-line_tools?language=objc @@ -54,13 +54,13 @@ }, { "type": "cppbuild", "label": "Build 02 chaining", - "command": "/opt/homebrew/opt/llvm/bin/clang++", + "command": "/usr/local/opt/llvm/bin/clang++", "args": [ "-std=c++17", "-stdlib=libc++", "-O2", // OpenMP includes & configuration - "-L/opt/homebrew/opt/libomp/lib", "-fopenmp", + "-L/usr/local/opt/libomp/lib", "-fopenmp", // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", @@ -81,18 +81,18 @@ "kind": "build", "isDefault": true }, - "detail": "compiler: /opt/homebrew/opt/llvm/bin/clang++", + "detail": "compiler: /usr/local/opt/llvm/bin/clang++", "dependsOn": ["Build .metallib 02"], }, { "type": "cppbuild", "label": "Build 02", - "command": "/opt/homebrew/opt/llvm/bin/clang++", + "command": "/usr/local/opt/llvm/bin/clang++", "args": [ "-std=c++17", "-stdlib=libc++", "-O2", // OpenMP includes & configuration - "-L/opt/homebrew/opt/libomp/lib", "-fopenmp", + "-L/usr/local/opt/libomp/lib", "-fopenmp", // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", @@ -113,18 +113,18 @@ "kind": "build", "isDefault": true }, - "detail": "compiler: /opt/homebrew/opt/llvm/bin/clang++", + "detail": "compiler: /usr/local/opt/llvm/bin/clang++", "dependsOn": ["Build .metallib 02"], }, { "type": "cppbuild", "label": "Build benchmark paper 02", - "command": "/opt/homebrew/opt/llvm/bin/clang++", + "command": "/usr/local/opt/llvm/bin/clang++", "args": [ "-std=c++17", "-stdlib=libc++", "-O2", // OpenMP includes & configuration - "-L/opt/homebrew/opt/libomp/lib", "-fopenmp", + "-L/usr/local/opt/libomp/lib", "-fopenmp", // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", @@ -145,7 +145,7 @@ "kind": "build", "isDefault": true }, - "detail": "compiler: /opt/homebrew/opt/llvm/bin/clang++", + "detail": "compiler: /usr/local/opt/llvm/bin/clang++", "dependsOn": ["Build .metallib 02"], }, { "type": "shell", @@ -169,13 +169,13 @@ }, { "type": "cppbuild", "label": "Build 03", - "command": "/opt/homebrew/opt/llvm/bin/clang++", + "command": "/usr/local/opt/llvm/bin/clang++", "args": [ "-std=c++17", "-stdlib=libc++", "-O2", // OpenMP includes & configuration - "-L/opt/homebrew/opt/libomp/lib", "-fopenmp", + "-L/usr/local/opt/libomp/lib", "-fopenmp", // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", @@ -196,19 +196,19 @@ "kind": "build", "isDefault": true }, - "detail": "compiler: /opt/homebrew/opt/llvm/bin/clang++", + "detail": "compiler: /usr/local/opt/llvm/bin/clang++", "dependsOn": ["Build .metallib 03"], }, { "type": "cppbuild", "label": "Build benchmark paper 03", - "command": "/opt/homebrew/opt/llvm/bin/clang++", + "command": "/usr/local/opt/llvm/bin/clang++", "args": [ "-std=c++17", // "-stdlib=libc++", "-std=c++11", "-O2", // OpenMP includes & configuration - "-L/opt/homebrew/opt/libomp/lib", "-fopenmp", + "-L/usr/local/opt/libomp/lib", "-fopenmp", // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", @@ -229,7 +229,7 @@ "kind": "build", "isDefault": true }, - "detail": "compiler: /opt/homebrew/opt/llvm/bin/clang++", + "detail": "compiler: /usr/local/opt/llvm/bin/clang++", "dependsOn": ["Build .metallib 03"], }, { "type": "shell", @@ -252,4 +252,4 @@ }, ] -} \ No newline at end of file +} From e8b45e4708a89fa8a2643bfee6f708cc558cd111 Mon Sep 17 00:00:00 2001 From: Will MacCormack Date: Fri, 20 Oct 2023 13:53:42 -0400 Subject: [PATCH 2/3] Make tasks.json json compliant --- .vscode/tasks.json | 28 +++++++--------------------- 1 file changed, 7 insertions(+), 21 deletions(-) diff --git a/.vscode/tasks.json b/.vscode/tasks.json index a10de11..969db48 100644 --- a/.vscode/tasks.json +++ b/.vscode/tasks.json @@ -8,9 +8,7 @@ "-std=c++17", "-stdlib=libc++", "-O2", - // OpenMP includes & configuration "-L/usr/local/opt/libomp/lib", "-fopenmp", - // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", "-g", @@ -29,9 +27,8 @@ "isDefault": true }, "detail": "compiler: /usr/local/opt/llvm/bin/clang++", - "dependsOn": ["Build .metallib 01"], + "dependsOn": ["Build .metallib 01"] }, - // See: https://developer.apple.com/documentation/metal/shader_libraries/building_a_library_with_metal_s_command-line_tools?language=objc { "type": "shell", "label": "Build .air 01", @@ -59,9 +56,7 @@ "-std=c++17", "-stdlib=libc++", "-O2", - // OpenMP includes & configuration "-L/usr/local/opt/libomp/lib", "-fopenmp", - // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", "-g", @@ -82,7 +77,7 @@ "isDefault": true }, "detail": "compiler: /usr/local/opt/llvm/bin/clang++", - "dependsOn": ["Build .metallib 02"], + "dependsOn": ["Build .metallib 02"] }, { "type": "cppbuild", "label": "Build 02", @@ -91,9 +86,7 @@ "-std=c++17", "-stdlib=libc++", "-O2", - // OpenMP includes & configuration "-L/usr/local/opt/libomp/lib", "-fopenmp", - // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", "-g", @@ -114,7 +107,7 @@ "isDefault": true }, "detail": "compiler: /usr/local/opt/llvm/bin/clang++", - "dependsOn": ["Build .metallib 02"], + "dependsOn": ["Build .metallib 02"] }, { "type": "cppbuild", "label": "Build benchmark paper 02", @@ -123,9 +116,7 @@ "-std=c++17", "-stdlib=libc++", "-O2", - // OpenMP includes & configuration "-L/usr/local/opt/libomp/lib", "-fopenmp", - // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", "-g", @@ -146,7 +137,7 @@ "isDefault": true }, "detail": "compiler: /usr/local/opt/llvm/bin/clang++", - "dependsOn": ["Build .metallib 02"], + "dependsOn": ["Build .metallib 02"] }, { "type": "shell", "label": "Build .air 02", @@ -174,9 +165,7 @@ "-std=c++17", "-stdlib=libc++", "-O2", - // OpenMP includes & configuration "-L/usr/local/opt/libomp/lib", "-fopenmp", - // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", "-g", @@ -197,19 +186,16 @@ "isDefault": true }, "detail": "compiler: /usr/local/opt/llvm/bin/clang++", - "dependsOn": ["Build .metallib 03"], + "dependsOn": ["Build .metallib 03"] }, { "type": "cppbuild", "label": "Build benchmark paper 03", "command": "/usr/local/opt/llvm/bin/clang++", "args": [ "-std=c++17", - // "-stdlib=libc++", "-std=c++11", "-O2", - // OpenMP includes & configuration "-L/usr/local/opt/libomp/lib", "-fopenmp", - // Metal includes & configuration "-I${workspaceFolder}/metal-cpp", "-fno-objc-arc", "-framework", "Metal", "-framework", "Foundation", "-framework", "MetalKit", "-g", @@ -230,7 +216,7 @@ "isDefault": true }, "detail": "compiler: /usr/local/opt/llvm/bin/clang++", - "dependsOn": ["Build .metallib 03"], + "dependsOn": ["Build .metallib 03"] }, { "type": "shell", "label": "Build .air 03", @@ -250,6 +236,6 @@ ], "dependsOn": ["Build .air 03"] - }, + } ] } From 961efd1d3e530bb73c59a29b80d5320459aa98ca Mon Sep 17 00:00:00 2001 From: Will MacCormack Date: Fri, 20 Oct 2023 15:15:56 -0400 Subject: [PATCH 3/3] New Launch Targets + all samples working --- .vscode/launch.json | 49 ++++++++++++++++++++++------- 02-GeneralArrayOperations/ops.metal | 20 ++++++------ 03-2DKernels/ops.metal | 18 +++++------ 3 files changed, 55 insertions(+), 32 deletions(-) diff --git a/.vscode/launch.json b/.vscode/launch.json index 34ea2ff..0bbf127 100644 --- a/.vscode/launch.json +++ b/.vscode/launch.json @@ -1,14 +1,5 @@ { - "configurations": [{ - "name": "Python: Current File", - "type": "python", - "request": "launch", - "program": "${file}", - "console": "integratedTerminal", - "justMyCode": true, - "cwd": "${workspaceFolder}/paper", - - }, + "configurations": [ { "name": "MetalAdder benchmark debugging", "type": "cppdbg", @@ -33,6 +24,42 @@ "externalConsole": false, "MIMode": "lldb", "preLaunchTask": "Build 02" + }, { + "name": "MetalOperations Chaining debugging", + "type": "cppdbg", + "request": "launch", + "program": "${workspaceFolder}/02-GeneralArrayOperations/chaining_benchmark.x", + "args": [], + "stopAtEntry": false, + "cwd": "${workspaceFolder}/02-GeneralArrayOperations", + "environment": [], + "externalConsole": false, + "MIMode": "lldb", + "preLaunchTask": "Build 02 chaining" + }, { + "name": "MetalOperations benchmark paper debugging", + "type": "cppdbg", + "request": "launch", + "program": "${workspaceFolder}/paper/paper_benchmark_02.x", + "args": [], + "stopAtEntry": false, + "cwd": "${workspaceFolder}/02-GeneralArrayOperations", + "environment": [], + "externalConsole": false, + "MIMode": "lldb", + "preLaunchTask": "Build benchmark paper 02" + }, { + "name": "2DKernels debugging", + "type": "cppdbg", + "request": "launch", + "program": "${workspaceFolder}/03-2DKernels/benchmark.x", + "args": [], + "stopAtEntry": false, + "cwd": "${workspaceFolder}/03-2DKernels", + "environment": [], + "externalConsole": false, + "MIMode": "lldb", + "preLaunchTask": "Build 03" } ] -} \ No newline at end of file +} diff --git a/02-GeneralArrayOperations/ops.metal b/02-GeneralArrayOperations/ops.metal index 3b62e23..3e864bd 100644 --- a/02-GeneralArrayOperations/ops.metal +++ b/02-GeneralArrayOperations/ops.metal @@ -63,8 +63,7 @@ kernel void inspector( uint quadgroup_index_in_threadgroup [[quadgroup_index_in_threadgroup]], uint quadgroups_per_threadgroup [[quadgroups_per_threadgroup]], uint simdgroup_index_in_threadgroup [[simdgroup_index_in_threadgroup]], - uint simdgroups_per_threadgroup [[simdgroups_per_threadgroup]], - uint thread_execution_width [[thread_execution_width]], + uint simdgroups_per_threadgroup [[simdgroups_per_threadgroup]], uint thread_index_in_quadgroup [[thread_index_in_quadgroup]], uint thread_index_in_simdgroup [[thread_index_in_simdgroup]], uint thread_index_in_threadgroup [[thread_index_in_threadgroup]], @@ -88,15 +87,14 @@ kernel void inspector( store[8] = quadgroups_per_threadgroup; store[9] = simdgroup_index_in_threadgroup; store[10] = simdgroups_per_threadgroup; - store[11] = thread_execution_width; - store[12] = thread_index_in_quadgroup; - store[13] = thread_index_in_simdgroup; - store[14] = thread_index_in_threadgroup; - store[15] = thread_position_in_threadgroup; - store[16] = threadgroup_position_in_grid; - store[17] = threadgroups_per_grid; - store[18] = threads_per_simdgroup; - store[19] = threads_per_threadgroup; + store[11] = thread_index_in_quadgroup; + store[12] = thread_index_in_simdgroup; + store[13] = thread_index_in_threadgroup; + store[14] = thread_position_in_threadgroup; + store[15] = threadgroup_position_in_grid; + store[16] = threadgroups_per_grid; + store[17] = threads_per_simdgroup; + store[18] = threads_per_threadgroup; } } diff --git a/03-2DKernels/ops.metal b/03-2DKernels/ops.metal index c637805..969bc42 100644 --- a/03-2DKernels/ops.metal +++ b/03-2DKernels/ops.metal @@ -133,7 +133,6 @@ kernel void inspector( uint quadgroups_per_threadgroup [[quadgroups_per_threadgroup]], uint simdgroup_index_in_threadgroup [[simdgroup_index_in_threadgroup]], uint simdgroups_per_threadgroup [[simdgroups_per_threadgroup]], - uint thread_execution_width [[thread_execution_width]], uint thread_index_in_quadgroup [[thread_index_in_quadgroup]], uint thread_index_in_simdgroup [[thread_index_in_simdgroup]], uint thread_index_in_threadgroup [[thread_index_in_threadgroup]], @@ -157,15 +156,14 @@ kernel void inspector( store[8] = quadgroups_per_threadgroup; store[9] = simdgroup_index_in_threadgroup; store[10] = simdgroups_per_threadgroup; - store[11] = thread_execution_width; - store[12] = thread_index_in_quadgroup; - store[13] = thread_index_in_simdgroup; - store[14] = thread_index_in_threadgroup; - store[15] = thread_position_in_threadgroup; - store[16] = threadgroup_position_in_grid; - store[17] = threadgroups_per_grid; - store[18] = threads_per_simdgroup; - store[19] = threads_per_threadgroup; + store[11] = thread_index_in_quadgroup; + store[12] = thread_index_in_simdgroup; + store[13] = thread_index_in_threadgroup; + store[14] = thread_position_in_threadgroup; + store[15] = threadgroup_position_in_grid; + store[16] = threadgroups_per_grid; + store[17] = threads_per_simdgroup; + store[18] = threads_per_threadgroup; } }