Skip to content
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

Jump to bottom

Separate host arch specific code to directories #75

Closed
wants to merge 1 commit into from
Closed

Separate host arch specific code to directories #75

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
11 changes: 11 additions & 0 deletions CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -78,6 +78,17 @@ include(${PROJECT_SOURCE_DIR}/cmake/dependencies.cmake)
add_library(umd_common_directories INTERFACE)
target_include_directories(umd_common_directories INTERFACE ${PROJECT_SOURCE_DIR} ${PROJECT_SOURCE_DIR}/device ${PROJECT_SOURCE_DIR}/third_party/fmt/include)

if(CMAKE_SYSTEM_PROCESSOR MATCHES "x86_64" OR CMAKE_SYSTEM_PROCESSOR MATCHES "i386")
Copy link
Contributor

Choose a reason for hiding this comment

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

Per offline discussion, include .cpp sources by host arch instead of headers, that way interface to metal will be the same.

The drawback, though, will be that the compiler won't be able to remove function call in that case. Wondering if that might even be a problem... According to github copilot, this shouldn't be a problem.

It could be a perf hit though.
In anycase, these should be in a separate utils repo. Started an issue here: #78

Copy link
Contributor Author

Choose a reason for hiding this comment

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

These function cannot be inline anymore since the implementation will move to .cpp files. Are we ok with this?

If we want to keep it inline, we could just guard include headers with defines like

#if defined (__x86_64__)
#include "x86/driver_atomics.h"
#endif

// empty file for device/driver_atomics.h

and move the implementation to separate .h files based on the architecture. This way we don't need to change cmake. Next step could be to add dependency on host arch to cmake (or investigate if there already is some dependency) in order to completely remove device/driver_atomics.h

message(STATUS "Building for host arch - x86")
target_include_directories(umd_common_directories INTERFACE ${PROJECT_SOURCE_DIR}/device/arch/x86)
elseif(CMAKE_SYSTEM_PROCESSOR MATCHES "arm" OR CMAKE_SYSTEM_PROCESSOR MATCHES "aarch64")
message(STATUS "Building UMD for host arch - ARM")
target_include_directories(umd_common_directories INTERFACE ${PROJECT_SOURCE_DIR}/device/arch/aarch64)
elseif(CMAKE_SYSTEM_PROCESSOR MATCHES "riscv")
message(STATUS "Building UMD for host arch - riscv")
target_include_directories(umd_common_directories INTERFACE ${PROJECT_SOURCE_DIR}/device/arch/riscv)
endif()

if(NOT DEFINED ENV{ARCH_NAME})
message(FATAL_ERROR "Please set ARCH_NAME to grayskull, wormhole_b0, or blackhole")
elseif($ENV{ARCH_NAME} STREQUAL "grayskull")
Expand Down
27 changes: 27 additions & 0 deletions device/arch/aarch64/driver_atomics.h
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,27 @@
/*
* SPDX-FileCopyrightText: (c) 2023 Tenstorrent Inc.
*
* SPDX-License-Identifier: Apache-2.0
*/

namespace tt_driver_atomics {

static inline __attribute__((always_inline)) void sfence() {
// Full memory barrier (full system). ARM does not have a Store-Any barrier.
// https://developer.arm.com/documentation/100941/0101/Barriers
asm volatile ("DMB SY" : : : "memory");
}

static inline __attribute__((always_inline)) void lfence() {
// Load-Any barrier (full system)
// https://developer.arm.com/documentation/100941/0101/Barriers
asm volatile ("DMB LD" : : : "memory");
}

static inline __attribute__((always_inline)) void mfence() {
// Full memory barrier (full system).
// https://developer.arm.com/documentation/100941/0101/Barriers
asm volatile ("DMB SY" : : : "memory");
}

} // namespace tt_driver_atomics
21 changes: 21 additions & 0 deletions device/arch/riscv/driver_atomics.h
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,21 @@
/*
* SPDX-FileCopyrightText: (c) 2023 Tenstorrent Inc.
*
* SPDX-License-Identifier: Apache-2.0
*/

namespace tt_driver_atomics {

static inline __attribute__((always_inline)) void sfence() {
asm volatile ("fence ow, ow" : : : "memory");
}

static inline __attribute__((always_inline)) void lfence() {
asm volatile ("fence ir, ir" : : : "memory");
}

static inline __attribute__((always_inline)) void mfence() {
asm volatile ("fence iorw, iorw" : : : "memory");
}

} // namespace tt_driver_atomics
24 changes: 24 additions & 0 deletions device/arch/x86/driver_atomics.h
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,24 @@
/*
* SPDX-FileCopyrightText: (c) 2023 Tenstorrent Inc.
*
* SPDX-License-Identifier: Apache-2.0
*/

#include <immintrin.h>

namespace tt_driver_atomics {

// Store-Any barrier.
static inline __attribute__((always_inline)) void sfence() {
_mm_sfence();
}
// Load-Any barrier.
static inline __attribute__((always_inline)) void lfence() {
_mm_lfence();
}
// Any-Any barrier.
static inline __attribute__((always_inline)) void mfence() {
_mm_mfence();
}

} // namespace tt_driver_atomics
2 changes: 1 addition & 1 deletion device/device_api_metal.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -6,4 +6,4 @@

#pragma once
#include "device/tt_device.h"
#include "device/driver_atomics.h"
#include "driver_atomics.h"
65 changes: 0 additions & 65 deletions device/driver_atomics.h
View file
Open in desktop

This file was deleted.

2 changes: 1 addition & 1 deletion device/simulation/deprecated/tt_versim_device.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5,7 +5,7 @@


#include "tt_device.h"
#include "device/driver_atomics.h"
#include "driver_atomics.h"
#include "common/logger.hpp"
#include <iostream>
#include <fstream>
Expand Down
2 changes: 1 addition & 1 deletion device/simulation/tt_simulation_device.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -11,7 +11,7 @@

#include "common/logger.hpp"
#include "common/assert.hpp"
#include "device/driver_atomics.h"
#include "driver_atomics.h"
#include "device/tt_cluster_descriptor.h"

#include "tt_simulation_device.h"
Expand Down
2 changes: 1 addition & 1 deletion device/tt_silicon_driver.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -58,7 +58,7 @@
#include <stdarg.h>
#include "device/cpuset_lib.hpp"
#include "common/logger.hpp"
#include "device/driver_atomics.h"
#include "driver_atomics.h"

#define WHT "\e[0;37m"
#define BLK "\e[0;30m"
Expand Down